即使使用 -arch=sm_20，Cuda Hello World printf 也无法正常工作答案

【问题标题】：Cuda Hello World printf not working even with -arch=sm_20即使使用 -arch=sm_20，Cuda Hello World printf 也无法正常工作
【发布时间】：2021-08-07 07:50:48
【问题描述】：

我不认为我是 Cuda 的新手，但显然我是。

我最近将我的 cuda 设备升级到了功能 1.3 到 2.1 (Geforce GT 630)。我还想对 Cuda 工具包 5.0 进行全面升级。

我可以编译通用 cuda 内核，但 printf 即使设置了 -arch=sm_20 也无法正常工作。

代码：

#include <stdio.h>
#include <assert.h>
#include <cuda.h>
#include <cuda_runtime.h>

__global__ void test(){

    printf("Hi Cuda World");
}

int main( int argc, char** argv )
{

    test<<<1,1>>>();
        return 0;
}

编译器：

Error   2   error MSB3721: The command ""C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v5.0\bin\nvcc.exe" -gencode=arch=compute_10,code=\"sm_20,compute_10\" --use-local-env --cl-version 2010 -ccbin "C:\Program Files (x86)\Microsoft Visual Studio 10.0\VC\bin"  -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v5.0\include" -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v5.0\include"  -G   --keep-dir "Debug" -maxrregcount=0  --machine 32 --compile -arch=sm_20  -g   -D_MBCS -Xcompiler "/EHsc /W3 /nologo /Od /Zi /RTC1 /MDd  " -o "Debug\main.cu.obj" "d:\userstore\documents\visual studio 2010\Projects\testCuda\testCuda\main.cu"" exited with code 2.  C:\Program Files (x86)\MSBuild\Microsoft.Cpp\v4.0\BuildCustomizations\CUDA 5.0.targets  592 10  testCuda
Error   1   error : calling a __host__ function("printf") from a __global__ function("test") is not allowed d:\userstore\documents\visual studio 2010\Projects\testCuda\testCuda\main.cu    9   1   testCuda

由于这个问题，我的生活即将结束……完成了。请在屋顶上告诉我答案。

【问题讨论】：

你有一组令人困惑的开关被传递给 nvcc：-gencode=arch=compute_10,code=\"sm_20,compute_10\" 和 -arch=sm_20。第一组可能与第二组冲突。您是如何指定（在项目设置中）为 arch sm_20 编译的？如果你编译 sm_10/compute_10 at all（即使你也指定了 sm_20），你会得到那个错误信息。因此，在您的项目设置中，您可能希望删除 sm_10/compute_10 代码生成。
您的项目正在尝试为 compute_10 和 compute_20 编译。不要为 compute_10 目标编译，它们在内核 printf 中不支持。
我也刚刚想通了。我的显卡有能力2.1，所以我想把compute_21，sm_21。当失败时，我认为这些值是完全独立的，所以我把它留在了 compute_10。此外，我没有看到之前讨论使用 printf 或其他需要 sm_20 的函数时提到的 compute_20。
有人随时回答，我可以选择你的答案。

标签： c++ cuda

【解决方案1】：

如果你在内核中使用printf，你应该使用cudaDeviceSynchronize()：

#include <stdio.h>
#include <assert.h>
#include <cuda.h>
#include <cuda_runtime.h>

__global__ void test(){
    printf("Hi Cuda World");
}

int main( int argc, char** argv )
{
    test<<<1,1>>>();
    cudaDeviceSynchronize();
    return 0;
}

【讨论】：

【解决方案2】：

在内核中 printf 仅在计算能力 2 或更高的硬件中受支持。因为您的项目设置为为 both 计算能力 1.0 和计算 2.1 构建，所以 nvcc 会多次编译代码并构建多架构 fatbinary 对象。错误是在计算能力 1.0 编译周期产生的，因为该架构不支持 printf 调用。

如果您从项目中删除计算能力 1.0 构建目标，错误将消失。

你也可以这样写内核：

__global__ void test()
{
#if __CUDA_ARCH__ >= 200
    printf("Hi Cuda World");
#endif
}

在为计算能力 2.0 或高目标构建时，__CUDA_ARCH__ 符号将仅 >= 200，这将允许您为计算能力 1.x 设备编译此代码而不会遇到语法错误。

当为正确的架构编译并且没有输出时，您还需要确保内核完成并且驱动程序刷新输出缓冲区。为此，在主机代码中内核启动后添加一个同步调用

例如：

int main( int argc, char** argv )
{

    test<<<1,1>>>();
    cudaDeviceSynchronize();
    return 0;
}

[免责声明：所有代码在浏览器中编写，从未编译，使用风险自负]

如果你同时做这两件事，你应该能够编译、运行并查看输出。

【讨论】：

【解决方案3】：

只需使用cudaDeviceSynchronize()。作为@Tomasz 答案的补充。

具有 2.x 或更高计算能力的设备支持从 CUDA 内核中调用 printf。

printf 输出存储在固定大小的循环缓冲区中。此缓冲区仅被刷新用于：

内核启动开始
同步（例如 cudaDeviceSynchronize()）
阻塞内存拷贝（例如 cudaMemcpy(...)）
模块加载/卸载
上下文破坏

所以最简单的“Hello world”例子：

#include <stdio.h>

__global__ void hello() {
    printf("Hello from GPU);
}

int main() {
    hello<<<1, 1>>>();
    cudaDeviceSynchronize();
}

参考：

【讨论】：