cuda - 有效的最小 GPU 线程数

标签 cuda gpu

我将在 CUDA 上并行化一些优化问题的本地搜索算法。这个问题非常难,所以实际可解决的问题的规模很小。
我担心的是,计划在一个内核中运行的线程数量不足以在 GPU 上获得任何加速(即使假设所有线程都已合并,没有银行冲突,非分支等)。
假设为 100 个线程启动了一个内核。期望从使用 GPU 中获利是否合理?如果线程数是 1000 怎么办?分析案件需要哪些额外信息?

最佳答案

100 个线程是不够的。理想情况下,您需要一个可以划分为至少与 GPU 上的多处理器 (SM) 一样多的线程 block 的大小,否则您将让处理器处于空闲状态。出于同样的原因,每个线程 block 应该有不少于 32 个线程。理想情况下,每个 block 应该有 32 个线程的小倍数(比如 96-512 个线程),如果可能,每个 SM 应该有多个这些 block 。

至少,您应该尝试有足够的线程来覆盖 SM 的算术延迟,这意味着在 Compute Capability 2.0 GPU 上,每个 SM 需要大约 10-16 个 warp(32 个线程组)。不过,它们并不都需要来自同一个线程 block 。这意味着,例如,在具有 14 个 SM 的 Tesla M2050 GPU 上,您将需要至少 4480 个线程,分成至少 14 个 block 。

也就是说,比这更少的线程也可以提供加速——这取决于许多因素。例如,如果计算受带宽限制,并且您可以将数据保存在设备内存中,那么您可以获得加速,因为 GPU 设备内存带宽高于 CPU 内存带宽。或者,如果它受计算限制,并且有很多指令级并行性(来自同一线程的独立指令),那么您将不需要那么多线程来隐藏延迟。后一点在 Vladimir Volkov 的 "Better performance at lower occupancy" talk 中得到了很好的描述。来自 GTC 2010。

主要是确保您使用所有的 SM:如果不这样做,您就没有使用 GPU 可以提供的所有计算性能或带宽。

关于cuda - 有效的最小 GPU 线程数,我们在Stack Overflow上找到一个类似的问题: https://stackoverflow.com/questions/7030063/

相关文章:

ffmpeg - 使用 FFmpeg 和 Nvidia GPU 加速

c++ - CUDA 扭曲和每个 block 的最佳线程数

cuda - 如何在没有分支发散的情况下检查 CUDA 内核中的数组边界

c++ - CUDA 在执行期间结合线程独立(??)变量

python - 如何判断 tensorflow 是否从 python shell 内部使用 gpu 加速?

python - 运行时错误 : CUDA error: CUBLAS_STATUS_EXECUTION_FAILED when calling `cublasSgemm( handle)` with GPU only

cuda - GPU从CPU读取还是CPU写入GPU?

c++ - cuBLAS cublasSgemv “Segmentation fault"

cuda - GPU 在 Julia 集合计算中没有提供性能改进

optimization - 在一维网格中计算扭曲ID/车道ID的最有效方法是什么?