为什么 printf() 在内核中可以工作，但使用 std::cout 却不能？

Question

我一直在探索并行编程领域，并用 Cuda 和 SYCL 编写了基本内核。我遇到过一种情况，我必须在内核内部打印，我注意到std::cout内核内部不起作用，而实际上却起作用printf。例如，考虑以下 SYCL 代码 - 这有效 -

void print(float*A, size_t N){
    buffer<float, 1> Buffer{A, {N}};
    queue Queue((intel_selector()));
    Queue.submit([&Buffer, N](handler& Handler){
       auto accessor = Buffer.get_access<access::mode::read>(Handler);
       Handler.parallel_for<dummyClass>(range<1>{N}, [accessor](id<1>idx){
           printf("%f", accessor[idx[0]]);
       });
    });
}

printf而如果我用它替换std::cout<<accessor[idx[0]]它会引发编译时错误，并提示 - Accessing non-const global variable is not allowed within SYCL device code. CUDA 内核也会发生类似的情况。这让我思考，两者之间可能存在什么差异printf，以及std::coout是什么导致了这种行为。

另外假设如果我想实现一个从GPU调用的自定义打印函数，我应该怎么做？
TIA

Answer 1

这让我思考 printf 和 std::cout 之间可能有什么区别导致这种行为。

是，有一点不同。printf()在你的内核中运行的不是标准C库printf()。对设备上函数进行不同的调用（其代码已关闭，如果它存在于 CUDA C 中）。该函数使用 NVIDIA GPU 上的硬件机制 - 用于内核线程打印的缓冲区，该缓冲区被发送回主机端，然后 CUDA 驱动程序将其转发到启动内核的进程的标准输出文件描述符。

std::cout没有得到这种编译器辅助的替换/劫持 - 并且它的代码与 GPU 完全无关。

不久前，我实现了一种std::cout用于 GPU 内核的类似机制；请参阅我在此处的回答以获取更多信息和链接。但是 - 我决定我不太喜欢它，而且它的编译相当昂贵，因此，我改编了printf()GPU 的 -family 实现，它现在是cuda-kat库（开发分支）的一部分。

这意味着我必须自己回答你的第二个问题：

如果我想实现一个从GPU调用的自定义打印函数，我应该怎么做？

printf()除非您有权访问未公开的 NVIDIA 内部结构 - 唯一的方法是在主机端使用调用而不是 C 标准库或系统调用。您本质上需要在低级原始 I/O 设施上模块化整个流。这绝非小事。