如何解决工作组中部分关闭的工人
让我们考虑一下,当 OpenCL 中有多个工作组和多个工作人员时。 如果我们在工作组中拥有与 GPU 上的“核心”一样多的工作人员,GPU 将按顺序在工作组中工作,其中工作组中的每个工作人员并行(对吗?)。完成一个工作组后,将执行下一个工作组。 如果我们在工作组中的工作人员比 GPU 上的“核心”少,据我所知,GPU 将并行执行多个工作组,其中导致多个工作人员并行执行(对吗?)。在这种情况下,当这段代码被执行时会发生什么?
__kernel void vector_add(__global const int *A)
{
// Get the index of the current element to be processed
int i = get_global_id(0);
if(A[i] == 0)
{
return; //stop like the half of all workers in a workgroup
}
// Do some time consuming calculation
...
}
此代码会导致“分支”,但工作组中的某些工作人员会直接停止(返回)其他人进行一些耗时的计算。我们可以称之为“分支”吗? 最大的问题是:返回的“核心”将做什么?他们是在等待工作组中的每个工人完成他的工作吗?或者因为很多人同时返回,他们会跳到下一个工作组执行吗?
行为供应商是特定的吗?或者这种情况是否在 OpenCL 中正确定义?
解决方法
如果你在内核和工作组中有分支,一些工作人员执行分支 A 和一些分支 B,所有工作人员必须计算两个分支并分别丢弃未使用的分支结果。这会对执行时间产生负面影响,这也是尽可能避免在 GPU 上进行分支的原因。在带有空 return
分支的示例中,如果工作组中只有一个工作人员必须执行 time consuming calculation
,则所有其他工作人员都必须等待,从而阻塞其他工作组的硬件资源。如果工作组很小并且您很幸运所有线程都执行 return
分支,那么该特定工作组的执行速度非常快。
物理GPU“核心”与工作组大小的匹配与计算结果无关,但会在一定程度上影响性能。工作组大小应该是 32 的倍数(GPU 将其“核心”细分为 32 个一组,即所谓的扭曲)。因此,如果工作组大小为 16,则一半的 GPU 将始终处于空闲状态。另一方面,如果工作组规模非常大(例如 1024)并且您在内核中有分支,那么所有工作人员都执行相同分支并最终出现上述情况的可能性较小。
如果您需要通过 local
内存在整个工作组中进行通信,有时工作组的大小需要进行权衡。更大的工作组允许更多的本地通信,但增加了“双分支”的可能性。如果您不使用 local
内存,您可以自由调整工作组大小以获得最佳性能(通常为 64-256)。
理想情况下,您希望用数百万个线程使 GPU 饱和,从而没有空闲的“内核”和最佳性能。
版权声明:本文内容由互联网用户自发贡献,该文观点与技术仅代表作者本人。本站仅提供信息存储空间服务,不拥有所有权,不承担相关法律责任。如发现本站有涉嫌侵权/违法违规的内容, 请发送邮件至 dio@foxmail.com 举报,一经查实,本站将立刻删除。