计算正确的工作组数量及其大小 OpenCL答案

【问题标题】：Calculating the right number of workgroups and their size OpenCL计算正确的工作组数量及其大小 OpenCL
【发布时间】：2016-01-15 19:34:33
【问题描述】：

我是 OpenCL 新手，我正在尝试理解这个由 Apple here 编写的示例程序。

程序的目标是计算输入数组中每个元素的平方，并将结果写入一个新数组。

您可以看到输入数组的维度为：1024。工作组的数量为 1024，每个工作组的大小为最大 CL_KERNEL_WORK_GROUP_SIZE。

如果内核中没有 get_local_id() 调用，谁能解释我在每个工作组中使用这么多工作项的意义何在？他们可以使用 1 作为每个工作组的大小吗？会有什么不同？

谢谢。

一些代码来说明这一点：

    // Get the maximum work group size for executing the kernel on the device
    //
    err = clGetKernelWorkGroupInfo(kernel, device_id, CL_KERNEL_WORK_GROUP_SIZE, sizeof(local), &local, NULL);


    // Execute the kernel over the entire range of our 1d input data set
    // using the maximum number of work group items for this device
    //
    global = count;
    err = clEnqueueNDRangeKernel(commands, kernel, 1, NULL, &global, &local, 0, NULL, NULL);

【问题讨论】：

这不是 CUDA，大小不是本地*全局的，只是全局的。全局大小很重要，内核代码运行了那么多次。本地大小仅在您想要共享有关正在运行的项目的信息并且必须划分全局大小时才有用。例如在项目 0-1023 之间共享一些数据。如果您根本不共享数据，那么大小并不是很重要，但由于开销减少和设备占用增加，因此首选更大的大小。
@DarkZeros 好的，我明白了，在这种情况下，全局大小很重要，所以 clEnqueueNDRangeKernel 调用中的 local_work_size 可以是 1 或 NULL，实际上什么都不会改变，对吧？
是的，就是这样。这就是为什么 OpenCL 允许本地大小为 NULL 的原因，因此驱动程序会选择它认为最适合性能的大小。

标签： c opencl

【解决方案1】：

您的全局工作大小以本地工作大小的块执行（理论上），如果您将本地工作组大小设置为 1，那么它将在每个本地工作组中仅执行 1 个线程。在 GPU 上，工作组与计算单元匹配 - 如果工作组大小为 1，则 1 个线程可能会占用整个计算单元。这真的非常非常慢

【讨论】：

因此，在 GPU 上使用每个工作组的 1 个工作项执行代码当然比在 CPU 上执行相同的代码要快，但是您仍然可以通过在工作组中做更多的工作来改进增加工作项的数量。这是正确的吗？
重点是：如果我在苹果的例子中增加工作项的数量，我不会在效率方面得到更好的结果，因为我没有使用本地内存。对吗？
每个工作组 1 个工作项可能比使用 CPU 慢。 GPU 有 32 个计算核心，因此每个工作组 1 个工作项可能意味着一次只能运行 32 个“线程”，这非常慢。增加工作组大小将大大提高性能 - gpus 使用每个工作组线程来运行 wave，这会填充 GPU 上的 SIMD 资源并允许您在合并访问中获取内存。如果每个内核只使用 1 个线程，那么这些资源的利用将严重不足
可能，但苹果获得的线程大小是我相信 that 内核的“最佳”线程大小。此外，重要的是要记住本地内存不是灵丹妙药 - 更多的本地内存可以减少内核占用率（活动波前的数量），这本身是一个令人困惑的统计数据（高占用率有利于内存限制内核，由于缓存，低占用率显然有利于算术界内核）。因此，更大的工作组大小（使用更多的本地内存）不一定更好。工作组的大小也有上限：nvidia 为 1024ish，amd 为 256/512ish
这是我需要的澄清。现在很清楚了。谢谢。