OpenCL: правильные результаты на CPU, а не на GPU: как правильно управлять памятью? - PullRequest
Новая
       92

OpenCL: правильные результаты на CPU, а не на GPU: как правильно управлять памятью?

2 голосов
/ 11 января 2012 
__kernel void CKmix(__global short* MCL, __global short* MPCL,__global short *C,  int S,  int B)
{       
    unsigned int i=get_global_id(0);
    unsigned int ii=get_global_id(1);
    MCL[i]+=MPCL[B*ii+i+C[ii]+S];
}

Ядро швы в порядке, оно успешно компилируется, и я получил правильные результаты, используя ЦП в качестве устройства, но это было, когда у меня был выпуск программы и воссоздание объектов памяти при каждом вызове ядра, что для моего Цель тестирования - около 16000 раз.

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

OpenCLProgram = clCreateProgramWithSource(hContext[Plat-1][Dev-1],11, OpenCLSource, NULL ,NULL);
clBuildProgram(OpenCLProgram, 0,NULL,NULL, NULL,NULL);
ocKernel = clCreateKernel(OpenCLProgram, "CKmix", NULL);

Это также успешно. Причина, по которой у меня есть двумерный массив контекстов, заключается в том, что я перебираю все платформы и устройства и позволяю пользователю выбирать платформу и устройство для использования. 

WorkSize[0]=SN;
WorkSize[1]=NF;  

PinnedCCL = clCreateBuffer(hContext[Plat-1][Dev-1], CL_MEM_READ_WRITE| CL_MEM_ALLOC_HOST_PTR, sizeof(short) *NF, NULL, NULL);
PinnedMCL = clCreateBuffer(hContext[Plat-1][Dev-1], CL_MEM_READ_WRITE | CL_MEM_ALLOC_HOST_PTR, sizeof(short) * Z*NF, NULL, NULL);
PinnedMO =  clCreateBuffer(hContext[Plat-1][Dev-1], CL_MEM_READ_WRITE | CL_MEM_ALLOC_HOST_PTR, sizeof(short) * Z,NULL, NULL);
PinnedMTEMP =  clCreateBuffer(hContext[Plat-1][Dev-1], CL_MEM_READ_WRITE | CL_MEM_ALLOC_HOST_PTR, sizeof(short) * Z,NULL, NULL);

DevComboCCL = clCreateBuffer(hContext[Plat-1][Dev-1], CL_MEM_READ_WRITE, sizeof(short) *NF, NULL, NULL);    
DevMappedMCL = clCreateBuffer(hContext[Plat-1][Dev-1], CL_MEM_READ_WRITE , sizeof(short) * Z*NF, NULL,NULL);
DevMO =  clCreateBuffer(hContext[Plat-1][Dev-1], CL_MEM_READ_WRITE , sizeof(short) * Z,NULL, NULL);

MO = (short*) clEnqueueMapBuffer(hCmdQueue[Plat-1][Dev-1], PinnedMO, CL_TRUE, CL_MAP_READ, 0, sizeof(short)*Z, 0, NULL, NULL, NULL);
CCL = (short*) clEnqueueMapBuffer(hCmdQueue[Plat-1][Dev-1], PinnedCCL, CL_TRUE, CL_MAP_WRITE, 0, sizeof(short)*NF, 0, NULL, NULL,NULL);
MCL = (short*) clEnqueueMapBuffer(hCmdQueue[Plat-1][Dev-1], PinnedMCL, CL_TRUE, CL_MAP_WRITE, 0, sizeof(short)*Z*NF, 0, NULL, NULL, NULL);
MTEMP = (short*) clEnqueueMapBuffer(hCmdQueue[Plat-1][Dev-1], PinnedMTEMP, CL_TRUE, CL_MAP_READ, 0, sizeof(short)*Z, 0, NULL, NULL, NULL);

for (n=0; n < Z; ++n) {
    MTEMP[n]=0;
    }

clSetKernelArg(ocKernel, 0, sizeof(cl_mem), (void*) &DevMO);
clSetKernelArg(ocKernel, 1, sizeof(cl_mem), (void*) &DevMCL);    
clSetKernelArg(ocKernel, 2, sizeof(cl_mem), (void*) &DevCCL);
clSetKernelArg(ocKernel, 3, sizeof(int),    (void*) &SH);
clSetKernelArg(ocKernel, 4, sizeof(int),    (void*) &SN);

Вышеизложенное составляет мою инициализацию, а остальное ниже повторяется. 

clEnqueueWriteBuffer(hCmdQueue[Plat-1][Dev-1], DevMCL, CL_TRUE, 0, Z*NF*sizeof(short), MCL, 0, NULL, NULL);
clEnqueueWriteBuffer(hCmdQueue[Plat-1][Dev-1], DevCCL, CL_TRUE, 0, NF*sizeof(short), CCL, 0, NULL, NULL);
clEnqueueWriteBuffer(hCmdQueue[Plat-1][Dev-1], DevMO, CL_TRUE, 0, Z*sizeof(short), MTEMP, 0, NULL, NULL);

clEnqueueNDRangeKernel(hCmdQueue[Plat-1][Dev-1], ocKernel, 2, NULL, WorkSize, NULL, 0, NULL, NULL);
clEnqueueReadBuffer(hCmdQueue[Plat-1][Dev-1],DevMO, CL_TRUE, 0, Z * sizeof(short),(void*) MO , 0, NULL, NULL);

Я проверил ошибки и не получаю никаких ошибок. Ядро многократно запускается с новыми данными. Я не уверен, где я делаю неправильно.

NVIDIA 550 ti compute 2.1, последний Dev Driver, Cuda SDK 4.0,

1 Ответ

2 голосов
/ 11 января 2012

Я не знаю, единственная ли это проблема с кодом, но это: 

unsigned int i=get_global_id(0);
unsigned int ii=get_global_id(1);
MCL[i]+=MPCL[B*ii+i+C[ii]+S];

определенно не очень хорошая идея.Обычно вы получаете несколько потоков, работающих с одним и тем же global_id(0), поэтому несколько потоков могут попытаться обновить MCL[i] одновременно (обратите внимание, что += не является атомарным).Я бы предположил, что для процессора недостаточно генерируемых потоков, чтобы продемонстрировать такое поведение в большинстве случаев, в то время как наличие тысяч потоков в графическом процессоре почти наверняка приведет к проблемам.сделать это, чтобы иметь только одномерный рабочий набор и для каждого потока накапливать все значения, которые идут в одну позицию: 

unsigned int i=get_global_id(0);
short accum = MCL[i]; //or 0, if thats the start
for(int ii = 0; ii < size; ++ii)
  accum += MPCL[B*ii+i+C[ii]+S];
MCL[i] = accum;

Конечно, это может быть или не быть возможным.Если это не исправление, вероятно, не все будет так просто.

...