CUDA Thrust 与原始内核相比如何？

Posted 2023-02-19

【中文标题】CUDA Thrust 与原始内核相比如何？

【英文标题】：How does CUDA Thrust compare to a raw kernel?

【发布时间】：2019-11-19 12:41:28

【问题描述】：

我是 GPU 编程的新手，我不确定什么会导致最高效的代码。使用 Thrust 与编写自定义内核并自己管理内存的优缺点是什么？

如果有助于详细说明我的目标是什么：我有一个大矩阵，对于每个值，我需要执行一些向量运算。我知道我需要动态并行来完成这项任务，并且目前有一个自定义内核来遍历将调用其他内核的矩阵。我正在考虑是否应将内核替换为 Thrust 调用（例如，thrust::for_each）和/或是否应在内核中使用 Thrust 进行向量操作。

【参考方案1】：

在过去的大约 12 个月里，我从主要编写 CUDA 内核转变为主要使用 Thrust，然后又回到主要编写 CUDA 内核。一般来说，编写自己的 CUDA 内核应该提供更好的原始性能，但在更简单的测试用例中，差异应该可以忽略不计。

Thrust 模仿 C++ STL，因此它具有许多与 STL 相同的优点和缺点。也就是说，它旨在以非常通用的方式对数据向量进行操作。从这个角度来看，Thrust 在某些方面比 CUDA 更好，但不应被视为一种万能的解决方案。 Thrust 的主要优势在于抽象和可移植性等领域；您不必考虑块大小，并且很容易编写同样适用于设备或主机上的数据的函子，而显然 CUDA 内核只能在设备内存上运行。它还有许多非常有用的算法；不必编写自己的归约或排序算法真是太好了，因为 Thrust 提供了这些非常有效的实现。但是在后台，您的数据访问模式可能不会轻易匹配 Thrust 的设计目标，并且 Thrust 往往会执行大量临时内存分配（这在性能环境中通常不好；您可以破解其内存管理模型来缓存这些临时分配，但我不建议实际这样做，只需编写内核并自己完全控制内存使用）。

我现在首选的工作模式是几乎所有事情都使用 CUDA，但针对特定算法（例如排序）、原型代码或我希望实现在主机上同样良好工作的代码深入研究 Thrust 算法或设备。

【讨论】：

关于仅将 Thrust 用于内存管理的想法？

@Richard 这就是我一直在做的事情，但是我遇到了一个讨厌的错误（我认为是在 CUDA 10.0 中？），尝试重用从初始大小显着增长的推力向量会导致分割当我将指针传递给 CUDA 内核时出错。我没有费心尝试调试问题，我放弃了向量并开始自己管理所有内存。无论如何，从长远来看，这是更可取的，我想使用异步设备传输，而这不是你可以用推力向量做的事情。如果我确定不需要数组增长，那么我仍然使用向量，它们很好

@Richard：我的想法：有人需要编写一个体面的内存管理库，无论是仅用于 CUDA 还是更一般地，它都不会束缚您编写代码的特定方式。即不是 STL 分配器，也不是需要推力设备向量等的东西。