OpenCL：CPU 上而不是 GPU 上的正确结果：如何正确管理内存？

Question

__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];
}

内核接缝正常，它编译成功，并且我使用 CPU 作为设备获得了正确的结果，但那是我每次调用内核时释放程序并重新创建内存对象的时候，这对我来说测试目的约为16000次。

我发布的代码就是我现在的位置，尝试使用固定内存和映射。

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

这也是成功的。我有一个 2d 上下文数组的原因是我遍历所有平台和设备并允许用户选择要使用的平台和设备。

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 计算能力 2.1， 最新的开发驱动程序， Cuda SDK 4.0，

Answer 1

我不知道它是否是代码的唯一问题，但这是：

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]（注意+= 不是原子的）。我假设对于 CPU，在大多数情况下没有足够的线程来显示这种行为，而在 gpu 上拥有数千个线程几乎肯定会导致问题。

做到这一点最合理的方法是只有一个一维工作集，并为每个线程累积所有值，这些值会转到一个位置：

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;

当然，这可能可行，也可能不可行。如果不是这样，修复可能不会那么简单。