当一个循环内启动大量内核时，OpenCL程序就会冻结。

Question

我有一个循环(大约10亿次迭代)来启动OpenCL内核。每个内核由一个线程执行，并执行一个非常简单的操作。问题是，在执行了数百万次迭代之后，代码冻结(停止)，程序根本没有终止。它冻结了对clFinish()的调用。程序并不总是在相同的迭代中冻结。

如果clFinish()每1000个迭代调用一次，而不是在每次迭代中被调用一次，那么问题就消失了，所以我觉得问题就在于，clFinish()等待内核的结束，但是内核在调用clFinish()之前(不知何故)就被杀死了。还请注意，当我在循环中插入许多printf()调用时，问题就消失了！

当我在CPU设备上执行程序时(在我的笔记本电脑上，我使用AMD SDK)，我遇到了这个问题，我也在一台带有Nvidia Fermi GPU的机器上遇到了这个问题(Nvidia SDK和驱动程序，AMD SDK也安装在这台机器上)。

我正在检查每个OpenCL API调用后的错误，但没有检测到错误。为了使代码清晰，我删除了错误检查。

我的问题：

• 他们是否不正确地使用了下面的OpenCL API？
• 如果同时发射大量的OpenCL内核，它们会有问题吗？

代码是由我们的一些工具自动生成的，所以请不要问我为什么要调用一个只有一个线程的内核(这是另一个问题，我知道这样的代码不利于性能)。我的目标是了解代码中的问题是什么，应该在理论上没有任何问题地运行。

主机代码：

/* OpenCL initialization.  */
/* ... */
cl_mem dev_acc = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(double), NULL, &err);

for (int h0 = 1; h0 <= ni; h0 += 1)
  for (int h2 = 0; h2 < nj; h2 += 1)
    for (int h5 = 0; h5 < h2 - 1; h5 += 1) {
          size_t global_work_size[1] = {1};
          size_t block_size[1] = {1};
          cl_kernel kernel2 = clCreateKernel(program, "kernel2", &err);
          clSetKernelArg(kernel2, 0, sizeof(cl_mem), (void *) &dev_acc);
          clEnqueueNDRangeKernel(queue, kernel2, 1, NULL, global_work_size,block_size,0, NULL, NULL);
          clFinish(queue);
          clReleaseKernel(kernel2);
       }

内核代码：

__kernel void kernel2(__global double *acc)
{
   *acc = 1;
}

编译: gcc -O3 -lm -std=gnu99 polybench.c .c ocl_utilities.c symm_host.c -lOpenCL -lm -I/opt/AMDAPP/include -L/opt/AMDAPP/lib/x86_64

我使用的是Ubuntu12.04，内核3.2.0-29-泛型，X86_64，内存:2GB

Answer 1

看你的代码我都不知道从哪里开始..。

但是，如果它达到了OpenCL标准，它应该运行得很好。如果您正在使用的库的实现能够处理这个问题，这就是问题所在。

您应该做的第一件事是检查每个OpenCL API调用的错误代码。我想你是在“填满”你的命令队列，从OpenCL的警车里听到没有人听到的尖叫声。如果您使用clFinish，队列会不时地被清空，也许可以防止这种“过度填充”。

其他一些事情:一个内核是否真的是你想要的？OpenCL被设计为在SIMD体系结构上执行，这意味着单指令多个数据。因此，当大量线程对不同的数据执行相同的代码时，OpenCL的性能最好。

您不必每次在循环中创建内核：

size_t global_work_size[1] = {1};
size_t block_size[1] = {1};
cl_kernel kernel2 = clCreateKernel(program, "kernel2", &err);
for (int h0 = 1; h0 <= ni; h0 += 1)
  for (int h2 = 0; h2 < nj; h2 += 1)
    for (int h5 = 0; h5 < h2 - 1; h5 += 1) {
          clSetKernelArg(kernel2, 0, sizeof(cl_mem), (void *) &dev_acc);
          clEnqueueNDRangeKernel(queue, kernel2, 1, NULL, global_work_size,block_size,0, NULL, NULL);
          clFinish(queue);

       }
clReleaseKernel(kernel2);

最后一件事是，您的执行模式实际上只有一个线程：

如果可能的话，可以尝试这样的方法(我不知道您对内存的要求等等)：

cl_mem dev_acc = clCreateBuffer(context, CL_MEM_READ_WRITE, ni * nj * sizeof(double), NULL, &err);

size_t global_work_size[1];
global_work_size[0] = ni;
global_work_size[1] = nj;
size_t block_size[1] = {1};
cl_kernel kernel2 = clCreateKernel(program, "kernel2", &err);

// some loop
clSetKernelArg(kernel2, 0, sizeof(cl_mem), (void *) &dev_acc);
clSetKernelArg(kernel2, 1, sizeof(int), &h2);
clEnqueueNDRangeKernel(queue, kernel2, 1, NULL, global_work_size,block_size,0, NULL, NULL);

其中一个内核看起来有点像：

__kernel void kernel2(__global double *acc, int h5)
{
   int h0 = get_global_id(0);
   int h2 = get_global_id(1);
   int ni = get_global_size(0);
   int nj = get_global_size(1);
   // do stuff with ni, nj, h0, h2
   if (h5 < h2)
   {
      *acc = 1;
   }
}

Answer 2

kronos有一些很好的反馈，只是为了添加更多的输入：

1. 一个内核可以加载和编译一次，并在每个设备上使用多次。编译需要一些时间，所以最好在一开始就做一次。
2. 在OpenCL中启动时，最好将1个线程视为for循环的一个迭代。
3. 要处理遍历for循环的迭代，请在1、2或3维中使用clEnqueueNDRangeKernel(...)。
4. I always treat my OpenCL memory as 3D, since it is easier to convert 3D to 1D than it is 1D to 3D (for me at least). Please see my linked answer.