我们可以在 CUDA 内核中使用 printf 或任何其他类似的函数吗？

Question

我的矩阵加法示例：

 __global__ void matrix_add(float *a, float*b, float *c, int N)
{
    int index;
    int Row = blockIdx.y * blockDim.y + threadIdx.y;
    int Col = blockIdx.x * blockDim.x + threadIdx.x;

    int index = Row * N + Col;      // N is the order of the square matrix

    cd[index]= ad[index] + bd[index];

}

我可以在上述内核中使用 printf 或任何其他类似的函数吗？这样我就不需要将数据从设备传输到主机内存（即

cudaMemcpyDeviceToHost

）。如果是的话怎么办？如果没有那为什么不呢？

Answer 1

您可以使用 printf(..) 但仅适用于 cc2.x 或更高版本。
您可以在 CUDA 编程指南附录 B.16 中阅读更多相关信息。

Answer 2

显示内核结果而不导致数据复制回主机的唯一方法是使用其中一种图形互操作模式。 CUDA 支持 OpenGL 和 Direct3D 互操作性。有关如何使用这些的示例，请参阅 CUDA 编程指南。

__device__ printf()

（计算能力 >= 2.0）和

__device__ cuPrintf()

（计算能力 < 2.0), both cause implicit copying of the printed strings back to the host. Very probably, both of these also cause implicit serialization of all kernels that try to print at the same time, thus are typically used only for debugging.

如果您在调试器中运行 CUDA 应用程序，您在调试器中查看的设备值也已隐式复制到主机。

从您的问题中不清楚您是想避免将值复制回主机，还是只想避免“显式”复制这些值。如果是后者，那么 __device__ printf() 方法对于在主机上显示少量结果是可行的。避免显式复制值的另一种方法是使用

thrust::device_vector

。 Thrust是CUDA自带的一个库。它的灵感来自于 C++ STL。您可以在主机端读取和写入

device_vector

，并在后台与设备执行隐式复制。您还可以通过使用所谓的映射内存来引发隐式复制。通过映射内存，CUDA 硬件可以根据内核需要在主机和设备之间执行隐式内存复制。

这一切的原因是主机和设备之间的复制非常昂贵。通常，它们占用了总计算时间的很大一部分。因此，有必要仔细考虑这些副本何时以及如何发生。我提到的所有技术都有不同的性能影响，并且如何最好地处理复制是特定于应用程序的。

Answer 3

printf()

之外，您还可以使用

这个独立 printf 库

的CUDA 支持分支。这样做有以下好处：

更完整的格式字符串功能集：更多格式说明符和更多支持的选项。

sprintf()
• 的可用性，因此您可以将打印编组到缓冲区中，而不必在不同时间发出单独的 printf() ，从而增加其他线程的打印输出中断的风险。使用
vprintf()
• 和

  vsprintf()

  。通过闭包支持行为定制：用户提供的单字符输出函数和传递给它的不透明
void *extra_arg
• 。
但它也有缺点：

对您的项目有额外的依赖；您需要下载、构建和安装或以其他方式使该库可用于您的项目。并且您需要仔细地将编译的设备端代码链接到它。

• 比仅使用内置的构建时间更长
printf()
• 虽然该库在各种设备上广泛使用，但在 GPU 上并未广泛使用。
适当披露：我改编了这个库（Marco Paland 原创）以在 CUDA 中使用，并维护了 CPU 端版本。所以，我对它有偏见。