不支持 CUDA 外部调用答案

【问题标题】：CUDA External calls not supported不支持 CUDA 外部调用
【发布时间】：2011-08-25 00:57:13
【问题描述】：

我正在开发一个在 Fermi 卡上运行的 CUDA 4.0 应用程序。根据规范，Fermi 具有 Compute Capability 2.0，因此应该支持非内联函数调用。

我使用 nvcc 4.0 在一个不同的 obj 文件中编译我拥有的每个类。然后，我将它们全部用 g++-4.4 链接起来。

考虑以下代码：

[文件 A.cuh]

#include <cuda_runtime.h>

struct A
{
    __device__ __host__ void functionA();
};

[文件 B.cuh]

#include <cuda_runtime.h>

struct B
{
    __device__ __host__ void functionB();
};

[文件 A.cu]

#include "A.cuh"
#include "B.cuh"

void A::functionA()
{
    B b;
    b.functionB();
}

尝试用nvcc -o A.o -c A.cu -arch=sm_20 编译A.cu 输出Error: External calls are not supported (found non-inlined call to _ZN1B9functionBEv)。

我一定做错了什么，但是怎么办？

【问题讨论】：

标签： cuda

【解决方案1】：

functionB 未声明，因此被视为外部调用。正如错误所说，不支持外部调用。实现functionB，它会起作用。

【讨论】：

【解决方案2】：

正如this thread on the NVidia forums 中解释的那样，似乎即使 Fermi 支持非内联函数，nvcc 仍然需要在编译期间拥有所有可用的函数，即在同一个源文件中：没有链接器（是的，那是可惜……）。

【讨论】：

【解决方案3】：

没错，CUDA 5.0 做到了。我无法让它公开外部设备变量，但设备方法工作得很好。 默认情况下不是。

nvcc 选项是“-rdc=true”。在 Visual Studio 和 Nsight 中，它是 Configuration Properties -> CUDA C/C++ -> Common -> Generate Relocatable Device Code 下项目属性中的一个选项。

【讨论】：