主机循环的CUDA布尔变量

CUDA boolean variable for host loop

本文关键字：布尔 变量 CUDA 循环 主机      更新时间：2023-10-16

我的代码的一个非常简化的版本看起来像：

do {
    //reset loop variable b to 0/false
    b = 0;
    // execute kernel
    kernel<<<...>>>(b);
   // use the value of b for while condition
} while(b);

布尔变量b可以由kernel中的任何线程设置为true，它告诉我们是否继续运行循环。

使用cudaMalloc、cudaMemset和cudaMemcpy，我们可以创建/设置/复制device内存来实现这一点。然而，我只是发现了钉扎记忆的存在。使用cudaMalloHost分配b，并在内核之后立即调用cudaDeviceSynchronize，在一个简单的测试程序中提供了相当大的速度（~50%）。

固定内存是这个布尔变量b的最佳选择，还是有更好的选择？

您还没有显示您的初始代码和修改的代码，因此没有人知道您在帖子中所说的改进细节。

您的问题的答案因而异

  1. b是读取和写入的，或者仅在GPU内核内部写入。如果在缓存中未找到导致延迟的b，则读取可能需要直接从主机端获取实际值。另一方面，如果有进一步的操作可以使线程保持繁忙，那么写入的延迟可以被覆盖
  2. 修改值的频率。如果您经常在程序中访问该值，GPU可能会将变量保留在L2中，以避免主机端访问
  3. 访问b之间的内存操作频率。如果在访问b之间有许多内存事务，则缓存中的b更有可能被其他内容替换。因此，当再次访问时，在缓存中找不到b，因此需要进行耗时的主机访问

在主机端有b导致许多主机内存事务的情况下，将其保留在GPU全局内存中并在每次循环迭代结束时将其传输回是合乎逻辑的。您可以在与内核的流相同的流中使用异步副本来快速完成此操作，然后立即与主机同步。

以上所有项目均适用于启用缓存的设备。如果你的设备是pr费米（CC<2.0），情况就不同了。

相关文章：