Почему вложенный цикл OpenCL работает только для некоторых элементов

Question

Я пытаюсь реализовать следующий цикл в ядре 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
    , &status );
    cl_mem deviceInputs = clCreateBuffer( context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, M*sizeof(float), inputs
    , &status );
    cl_mem deviceWeights = clCreateBuffer( context, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, N*M*sizeof(float), weights
    , &status );

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

    status = clSetKernelArg( kernel, 0, sizeof(deviceGradients), &deviceGradients );
    status = clSetKernelArg( kernel, 1, sizeof(deviceInputs), &deviceInputs );
    status = clSetKernelArg( kernel, 2, sizeof(deviceWeights), &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;
    }

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

Answer 1

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

Глобальный размер работы указывает общее количество выполненных вызовов (рабочих элементов). 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);