13

您好：在 OpenCL 中，全局工作大小（维度）是否需要是工作组大小（维度）的倍数？

如果是这样，是否有处理矩阵而不是工作组维度的倍数的标准方法？我能想到两种可能：

将工作组维度的大小动态设置为全局工作维度的一个因子。（这会产生寻找因素的开销，并可能将工作组设置为非最佳大小。）

将全局工作的维度增加到工作组维度的最接近倍数，保持所有输入和输出缓冲区相同，但检查内核中的边界以避免段错误，即对超出所需输出范围的工作项不做任何事情. （这似乎是更好的方法。）

第二种方法可行吗？有没有更好的办法？（或者因为工作组维度不需要划分全局工作维度，所以没有必要？）

谢谢！

4

3 回答 3

7

谢谢乍得的链接。但实际上，如果您继续阅读：

如果指定了 local_work_size，则 global_work_size[0], … global_work_size[work_dim - 1] 中指定的值必须能被 local_work_size[0], … local_work_size[work_dim – 1] 中指定的相应值整除。

所以是的，本地工作大小必须是全局工作大小的倍数。

我还认为将全局工作大小分配给最接近的倍数并注意边界应该是可行的，当我有时间尝试它时，我会发表评论。

于 2010-07-02T10:53:41.383 回答
4

这似乎是一个旧帖子，但让我用一些新信息更新这个帖子。希望它可以帮助其他人。

全局工作大小（维度）是否需要是 OpenCL 中工作组大小（维度）的倍数？

答：在 OpenCL 2.0 之前都是如此。在CL2.0之前，你的全局work size必须是local work size的倍数，否则你在执行clEnqueueNDRangeKernel时会报错。

但是从 CL2.0 开始，这不再是必需的了。您可以使用适合您的应用程序尺寸的任何全局工作大小。但是，请记住，硬件实现可能仍然使用“旧”方式，这意味着填充全局工作组大小。因此，它使性能高度依赖于硬件架构。您可能会在不同的硬件/平台上看到完全不同的性能。另外，您想让您的应用程序向后兼容，以支持仅支持 CL 至 1.2 版本的旧平台。所以，我认为CL2.0中加入的这个新特性只是为了方便编程，为了获得更好的可控性能和向后兼容性，我建议你还是使用你提到的以下方法：

将全局工作的维度增加到工作组维度的最接近倍数，保持所有输入和输出缓冲区相同，但检查内核中的边界以避免段错误，即对超出所需输出范围的工作项不做任何事情. （这似乎是更好的方法。）

答：你完全正确。这是处理这种情况的正确方法。仔细设计本地工作组大小（考虑寄存器使用情况、缓存命中/未命中、内存访问模式等因素）。然后将您的全局工作大小填充到本地工作大小的倍数。然后，你可以走了。

要考虑的另一件事是，如果您的内核中有很多边界检查工作，您可以利用图像对象而不是缓冲区来存储数据。对于图像，边界检查由硬件自动完成，在大多数实现中几乎没有开销。因此，填充您的全局工作大小，将数据存储在图像对象中，然后，您只需要正常编写代码而无需担心边界检查。

于 2016-04-13T20:53:35.053 回答
1

根据标准，它不必来自我所看到的。我想我会用一个分支来处理它，但我不知道你在做什么类型的矩阵运算。

http://www.khronos.org/registry/cl/specs/opencl-1.1.pdf#page=131

global_work_size指向一个无符号值数组，这些值描述了将执行内核函数的维度work_dim中全局工作项的数量。work_dim全局工作项的总数计算为global_work_size[0]* ... * global_work_size[work_dim – 1]。

中指定的值 global_work_size+ 中指定的相应值global_work_offset 不能超过 sizeof(size_t)内核执行将在其上排队的设备给定的范围。sizeof(size_t)可以使用 CL_DEVICE_ADDRESS_BITS表 4.3确定设备的 例如，如果 CL_DEVICE_ADDRESS_BITS= 32，即设备使用 32 位地址空间，size_t则为 32 位无符号整数且global_work_size值必须在 1 .. 2^32 - 1 范围内。超出此范围的值将返回 CL_OUT_OF_RESOURCES错误。

于 2010-07-01T06:51:05.250 回答