Блог

Я пытаюсь реализовать следующий цикл в ядре OpenCL.

 for(i=0;i<N;i  ) for(j=0;j<M;j  ) weights[i*M j]  = gradients[i] * input[j];
  

Это мое ядро. В настоящее время я жестко кодирую M равным 4, и он работает только для первых 4 элементов.

 __kernel
void cwk3( __global float *gradients,  __global float *inputs,  __global float *weights)
{
    // The global id tells us the index of the vector for this thread.
    int gid1 = get_global_id(0);
    int gid2 = get_global_id(1);

    // Perform the addition.
    weights[(gid1 * 4)   gid2]  = gradients[gid1] * inputs[gid2];
}
  

Соответствующий код c

     float
        *gradients = (float*) malloc( N  *sizeof(float) ),
        *inputs    = (float*) malloc(   M*sizeof(float) ),
        *weights   = (float*) malloc( N*M*sizeof(float) );
    initialiseArrays( gradients, inputs, weights, N, M );

    cl_mem deviceGradients = clCreateBuffer( context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, N*sizeof(float), gradients
    , amp;status );
    cl_mem deviceInputs = clCreateBuffer( context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, M*sizeof(float), inputs
    , amp;status );
    cl_mem deviceWeights = clCreateBuffer( context, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, N*M*sizeof(float), weights
    , amp;status );

    cl_kernel kernel = compileKernelFromFile( "kernel.cl", "cwk3", context, device );

    status = clSetKernelArg( kernel, 0, sizeof(deviceGradients), amp;deviceGradients );
    status = clSetKernelArg( kernel, 1, sizeof(deviceInputs), amp;deviceInputs );
    status = clSetKernelArg( kernel, 2, sizeof(deviceWeights), amp;deviceWeights );

    size_t indexSpaceSize[2], workGroupSize[1];
    indexSpaceSize[0] = N;
    indexSpaceSize[1] = M;
    workGroupSize [0] = 4;

    status = clEnqueueNDRangeKernel( queue, kernel, 1, NULL, indexSpaceSize, workGroupSize, 0, NULL, NULL );
    if( status != CL_SUCCESS )
    {
        printf( "Failure enqueuing kernel: Error %d.n", status );
        return EXIT_FAILURE;        
    }

    status = clEnqueueReadBuffer( queue, deviceWeights, CL_TRUE, 0, N*M*sizeof(float), weights, 0, NULL, NULL );
    if( status != CL_SUCCESS )
    {
        printf( "Could not copy device data to host: Error %d.n", status );
        return EXIT_FAILURE;
    }
  

Это просто создает буферы и копирует их на графический процессор, запускает ядро, а затем считывает ответ с графического процессора на центральный процессор. N и M считываются в качестве аргументов командной строки. В настоящее время я устанавливаю для них обоих значение 4 для тестирования

Вы, кажется, запутались в глобальных и локальных рабочих группах.

Глобальный рабочий размер определяет общее количество выполненных вызовов (рабочих элементов). global_work_size=[M,N] вызовет ядро MxN раз в общей сложности. Один рабочий элемент может определять свою позицию с помощью get_global_id . OpenCL мог бы реализовать это как-то так :

 for(i=0;i<N;i  )
   for(j=0;j<M;j  )
       call_kernel(set global_id=[i,j])
  

Локальные рабочие группы описывают, как группировать отдельные запущенные рабочие элементы (которые создаются в соответствии с глобальными размерами) и информировать их друг о друге и совместно использовать память между собой. Ни одна из этих функций, которые вы используете / нуждаетесь, поэтому игнорируйте их.
Итак, чтобы реализовать ваш цикл for в OpenCL:

 for(i=0;i<N;i  ) 
    for(j=0;j<M;j  ) 
        weights[i*M j]  = gradients[i] * input[j];
  

У вас было бы это ядро:

  __kernel
void cwk3( __global float *gradients,  __global float *inputs,  __global float *weights)
{
    int gid1 = get_global_id(0);
    int gid2 = get_global_id(1);
    int M = get_global_size(0);

    weights[(gid1 * M)   gid2]  = gradients[gid1] * inputs[gid2];
}
  

И назовите это так:

 size_t global_work[2];
global_work[0]=M;
global_work[1]=N;
// This is 2D kernel, not 1D
// Offsets are 0
// Global work size is M*N
// Ignore local work size 
status = clEnqueueNDRangeKernel( queue, kernel, 2, NULL, global_work);