A partire da ora, la mia GPU è più lenta della mia CPU quando si tratta del tempo di esecuzione del kernel. Ho pensato che forse dal momento che stavo testando con un piccolo campione, la CPU ha finito per finire più velocemente a causa di un sovraccarico di avvio più piccolo. Tuttavia, quando ho testato il kernel con dati quasi 10 volte la dimensione del campione, la CPU stava ancora finendo più velocemente e la GPU era quasi 400ms dietro.Ottimizzazione del codice del kernel in opencl per una GPU
Runtime con il file 2.39MB CPU: 43.511ms GPU: 65.219ms
Runtime con file di 32.9MB CPU: 289.541ms GPU: 605.400ms
Ho provato ad utilizzare memoria locale, anche se io Sono sicuro al 100% che stavo usando male e ho avuto due problemi. Il kernel termina ovunque tra 1000-3000 ms (a seconda della dimensione che ho impostato per localWorkSize) o mi imbatto in un codice di stato di -5, che è CL_OUT_OF_RESOURCES.
Ecco il kernel che un altro membro SO mi ha aiutato.
__kernel void lowpass(__global float *Array, __global float *coefficients, __global float *Output) {
int globalId = get_global_id(0);
float sum=0.0f;
for (int i=0; i< 65; i++)
{
float tmp=0;
if (globalId+i > 63)
{
tmp=Array[i+globalId-64]*coefficients[64-i];
}
sum += tmp;
}
Output[globalId]=sum;
}
Questo è stato il mio tentativo di utilizzare la memoria locale. Il primo bit sarà uno snippet dal codice host e la seguente parte è il kernel.
//Set the size of localMem
status |= clSetKernelArg(
kernel,
2,
1024, //I had num_items*(float) but it gave me a -5. Num items is the amount of elements in my array (around 1.2 million elements)
null);
printf("Kernel Arg output status: %i \n", status);
//set a localWorkSize
localWorkSize[0] = 64;
//execute the kernel with localWorkSize included
status = clEnqueueNDRangeKernel(
cmdQueue,
kernel,
1,
NULL,
globalWorkSize,
localWorkSize,
0,
NULL,
&someEvent);
//Here is what I did to the kernel***************************************
__kernel void lowpass(__global float *Array, __global float *coefficients, __global float *Output, __local float *localMem) {
int globalId = get_global_id(0);
int localId = get_local_id(0);
localMem[localId] = globalId[globalId];
float sum=0.0f;
for (int i=0; i< 65; i++)
{
float tmp=0;
if (globalId+i > 63)
{
tmp=localMem[i+localId-64]*coefficients[64-i];
}
sum += tmp;
}
Output[globalId]=sum;
}
collegamento di riferimento che ho usato quando si cerca di impostare le variabili locali: How do I use local memory in OpenCL?
link utilizzato per trovare kernelWorkGroupSize (questo è il motivo per cui ho 1.024 set nel kernelArg): CL_OUT_OF_RESOURCES for 2 millions floats with 1GB VRAM?
I' Ho visto altre persone hanno problemi simili in cui la GPU è più lenta della CPU, ma per molti di loro usano clEnqueueKernel invece di clEnqueueNDRangeKernel.
Heres la mia domanda precedente, se avete bisogno di più informazioni su questo kernel: Best approach to FIFO implementation in a kernel OpenCL
Trovato alcuni trucchi di ottimizzazione per aswell di GPU. https://developer.amd.com/wordpress/media/2012/10/Optimizations-ImageConvolution1.pdf
Codice modificato; Errore esiste ancora
__kernel void lowpass2(__global float *Array, __global float *coefficients, __global float *Output) {
int globalId = get_global_id(0);
float sum=0.0f;
float tmp=0.0f;
for (int i=64-globalId; i< 65; i++)
{
tmp = 0.0f;
tmp=Array[i]*coefficients[i];
sum += tmp;
}
Output[globalId]=sum;
}
Sono abbastanza sicuro che * veramente * non si vuole uno 'if()' -statement nel tuo inner 'for'-loop. Un compilatore intelligente * potrebbe * essere in grado di sollevare "if" dal ciclo, ma un driver gpu * probabilmente * non ha il tempo o le capacità per farlo in modo efficiente. – EOF
Quale problema/algoritmo stai risolvendo/implementando? – mfa
@EOF Vado a dare un'occhiata alle istruzioni switch come alternativa a if(). – VedhaR