如何在设备内存中有效地混洗数据？

Question

题

在设备全局内存中移动许多随机(非合并)值时,这是最有效的方法吗？

注意:许多值大于500.

上下文

我已经在GPU的遗传算法实现中工作了一段时间,我一直在努力在我的框架的灵活性和GPU架构的微优化之间挣扎.GA数据始终驻留在GPU中.只有最佳世代解决方案才会复制到主机内存中.

详细方案

我正在优化迁移功能.这里基本上很少有数据在设备全局存储器中进行混洗.但我有在这样的方式我的数据以便它合并为GA运营商的内核线程的内存访问计划,这使得洗牌一对'的基因组 ’,迈进了单精度浮点数值,并与另一交换他们的事基因组中同样的跨越时尚.

已知的解决方案

问题不在于内存带宽,而在于调用延迟和线程阻塞使进程停滞.

1. 我写了几个设备内核,其功能仅仅是在地址之间移动值.这将启动一个内核(具有非常低的占用率,不同的代码和随机内存访问...因此它运行的小代码,将被序列化),但只能对设备进行两次内核调用.

   • 1st Kernel将值复制到缓冲区数组.
   • 第二个内核交换值.

2. 我知道我可以为每个值使用cudaMemcpy,但这需要多次调用cudaMemCpy,我认为这是同步调用.

简化的代码示例:

int needed_genome_idx = 0; // Some random index.
for(int nth_gene = 0; nth_gene < num_genes; ++nthgene)
{
  cudaMemcpy(genomes_buffer + nth_gene,
             src + needed_genome_idx + nth_gene * stride_size, // stride_size being a big number, usually equal to the size of the GA population.
             sizeof(float),
             cudaMemCpyDeviceToDevice);
}

这是一个可行的解决方案？使用cudaMemCpyAsync会有助于提高性能吗？

有没有更好的方法,或者至少更优雅的方式来做这样的记忆操作？

Answer 1

你可以尝试写一个kernel来完成shuffle，也许比多次调用cudaMemcpy更有效率。