sum - OpenCL 2.x - 减和函数
问题描述
从上一篇文章:strategy-for-doing-final-reduction,我想知道OpenCL 2.x 提供的最后一个功能(不是 1.x,这是上一篇文章的主题),尤其是关于原子函数它允许执行数组的缩减(在我的情况下是总和缩减)。
有人告诉我 OpenCL 1.x 原子函数 ( atom_add
) 的性能很差,我可以检查一下,所以我正在寻找一种方法来final reduction function
为)。
我回想起我目前使用的典型内核代码:
__kernel void sumGPU ( __global const double *input,
__global double *partialSums,
__local double *localSums)
{
uint local_id = get_local_id(0);
uint group_size = get_local_size(0);
// Copy from global memory to local memory
localSums[local_id] = input[get_global_id(0)];
// Loop for computing localSums
for (uint stride = group_size/2; stride>0; stride /=2)
{
// Waiting for each 2x2 addition into given workgroup
barrier(CLK_LOCAL_MEM_FENCE);
// Divide WorkGroup into 2 parts and add elements 2 by 2
// between local_id and local_id + stride
if (local_id < stride)
localSums[local_id] += localSums[local_id + stride];
}
// Write result into partialSums[nWorkGroups]
if (local_id == 0)
partialSums[get_group_id(0)] = localSums[0];
}
如您所见,在内核代码执行结束时,我得到了partialSums[number_of_workgroups]
包含所有部分和的数组。
您能否告诉我如何使用 OpenCL 2.x 提供的函数的最佳性能来执行该数组的第二次和最后一次缩减。一个经典的解决方案是使用 CPU 执行此最终缩减,但理想情况下,我想直接使用内核代码执行此操作。
欢迎提供代码片段的建议。
最后一点,我正在使用以下型号开发 MacOS High Sierra 10.13.5:
OpenCL 2.x 可以安装在我的硬件 MacOS 型号上吗?
解决方案
应避免使用原子函数,因为与并行缩减内核相比,它们会损害性能。你的内核看起来是在正确的轨道上,但你需要记住你必须多次调用它;不要在主机上执行最终的求和(除非您从之前的缩减中获得了非常少量的数据)。也就是说,您需要继续调用它,直到您的本地大小等于您的全局大小。由于无法在工作组之间进行同步,因此无法对大量数据进行单次调用。
此外,您要小心设置适当的工作组大小(即本地大小),这取决于本地和全局内存吞吐量和延迟。不幸的是,据我所知,除了自我分析代码之外,没有办法通过 OpenCL 来确定这一点,尽管编写起来并不难,因为 OCL 为您提供了 JIT 编译。通过经验测试,我发现您应该在遭受太多银行冲突(本地大小太大)与全局内存延迟损失(本地大小太小)之间找到一个最佳点。最好先做一个基准来确定减少的最佳局部大小,然后使用该局部大小进行未来的减少。
编辑:还值得注意的是,将内核调用链接在一起的最佳方法是通过 OpenCL 事件。
推荐阅读
- wordpress - Wordpress 经典编辑器正在删除图像标题
- unix - 如何使用 tcsh 在别名中使用多个 argv 元素?
- node.js - passport.serializeUser 错误如何解决?
- c++ - 尝试一个问题,但我编写的代码没有给出大量的输出。为什么?
- nvcc - 如何在 nvidia/cuda 基础 docker 中仅安装 nvcc?
- vba - 使用已经存在的名称创建类方法
- javascript - 动态注入具有脚本的 jquery html() 的潜在内存问题
- html - 为什么这个元素没有出现在 iPhone 上的所有浏览器中?
- python-3.x - 如何使用 python 将数据发送到扭曲类之外的服务器
- mysql - 将 MYSQL RAW 包含的哈希(加密)查询转换为 Laravel 雄辩的查询