我使用的是 AMD Radeon HD 7700 GPU。我想使用以下内核来验证波前尺寸是否为 64。
__kernel
void kernel__test_warpsize(
__global T* dataSet,
uint size
)
{
size_t idx = get_global_id(0);
T value = dataSet[idx];
if (idx<size-1)
dataSet[idx+1] = value;
}
在主程序中，我传递了一个包含 128 个元素的数组。初始值为dataSet[i]=i。在内核之后，我期望以下值：
数据集[0]=0
数据集[1]=0
数据集[2]=1
...
数据集[63]=62
数据集[64]=63
数据集[65]=63
数据集[66]=65
...
数据集[127]=126
但是，我发现dataSet[65]是64，而不是63，这不符合我的预期。
我的理解是，第一个波前（64个线程）应该将dataSet[64]更改为63。因此，当执行第二个波前时，线程#64应该获取63并将其写入dataSet[65]。但我看到 dataSet[65] 仍然是 64。为什么？
您正在调用未定义的行为。如果您希望访问工作组中另一个线程正在写入的内存，则必须使用屏障。
此外，假设 GPU 同时运行 2 个波前。那么 dataSet[65] 确实包含正确的值，第一个波前根本还没有完成。
此外，根据规范，所有项目的输出为 0 也是有效结果。这是因为一切也可以完全串行执行。这就是为什么你需要障碍。
根据您的评论我编辑了这部分：
Install http://developer.amd.com/tools-and-sdks/heterogeneous-computing/codexl/ http://developer.amd.com/tools-and-sdks/heterogeneous-computing/codexl/
Read: http://developer.amd.com/download/AMD_Accelerated_Parallel_Processing_OpenCL_Programming_Guide.pdf http://developer.amd.com/download/AMD_Accelerated_Parallel_Processing_OpenCL_Programming_Guide.pdf
优化一定数量线程内的分支只是优化的一小部分。您应该了解 AMD 硬件如何在工作组内调度波前，以及如何通过交错执行波前（在工作组内）来隐藏内存延迟。分支还会影响整个工作组的执行，因为运行它的有效时间基本上与执行单个运行时间最长的波前的时间相同（在组中的所有内容完成之前它无法释放本地内存等，因此它无法调度另一个工作组）。但这也取决于您的本地内存和寄存器使用情况等。要查看实际发生的情况，只需获取 CodeXL 并运行 GPU 分析即可。这将准确显示设备上发生的情况。
即使这仅适用于当前一代的硬件。这就是为什么这个概念不在 OpenCL 规范本身中的原因。这些属性变化很大，并且很大程度上取决于硬件。
但如果您真的想知道 AMD 波前尺寸是多少，答案几乎总是 64（请参阅http://devgurus.amd.com/thread/159153 http://devgurus.amd.com/thread/159153参考他们的 OpenCL 编程指南）。构成当前整个产品线的所有 GCN 设备均为 64。也许一些较旧的设备有 16 或 32，但现在所有设备都只有 64（对于 nvidia，一般是 32）。
本文内容由网友自发贡献，版权归原作者所有，本站不承担相应法律责任。如您发现有涉嫌抄袭侵权的内容，请联系:hwhale#tublm.com(使用前将#替换为@)
