当循环内启动大量内核时，OpenCL 程序冻结

Question

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

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

当我在 CPU 设备上（在我的笔记本电脑上，我使用 AMD SDK）上执行程序时遇到问题，并且我在带有 Nvidia Fermi GPU 的机器上也遇到问题（Nvidia SDK 和驱动程序，AMD SDK 也已安装在那台机器上）。

我在每次 OpenCL API 调用后检查错误，但未检测到错误。我删除了错误检查以使代码清晰。

我的问题：

• 他们是否对下面的 OpenCL API 有任何不正确的使用？

• 如果同时启动大量 OpenCL 内核会有什么问题吗？

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

主机代码：

/* 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 ocl_utilities.c symm_host.c -lOpenCL -lm -I/opt/AMDAPP/include -L/opt/AMDAPP/lib/x86_64

我使用的是 Ubuntu 12.04，内核 3.2.0-29-generic，X86_64，RAM：2 GB

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.