ptxas文件中的CUDA外部类链接和未解析的extern函数

Question

我正在使用CUDA，并创建了一个处理复数整数的int2_类。

ComplexTypes.h文件中的类声明如下：

namespace LibraryNameSpace
{
    class int2_ {

        public:
            int x;
            int y;

            // Constructors
            __host__ __device__ int2_(const int,const int);
            __host__ __device__ int2_();
            // etc.

            // Equalities with other types      
            __host__ __device__ const int2_& operator=(const int);
            __host__ __device__ const int2_& operator=(const float);
            // etc.

    };
}

ComplexTypes.cpp文件中的类实现如下：

#include "ComplexTypes.h"

__host__ __device__         LibraryNameSpace::int2_::int2_(const int x_,const int y_)           { x=x_; y=y_;}
__host__ __device__         LibraryNameSpace::int2_::int2_() {}
// etc.

__host__ __device__ const   LibraryNameSpace::int2_& LibraryNameSpace::int2_::operator=(const int a)                        { x = a;            y = 0.;             return *this; }
__host__ __device__ const   LibraryNameSpace::int2_& LibraryNameSpace::int2_::operator=(const float a)                      { x = (int)a;       y = 0.;             return *this; }
// etc.

一切都很好。在main (包括ComplexTypes.h)中，我可以处理int2_数字。

在CudaMatrix.cu文件中，我现在包括了ComplexTypes.h，并定义并正确实例化了__global__函数：

template <class T1, class T2>
__global__ void evaluation_matrix(T1* data_, T2* ob, int NumElements)
{
    const int i = blockDim.x * blockIdx.x + threadIdx.x;
    if(i < NumElements) data_[i] = ob[i];
}

template __global__ void evaluation_matrix(LibraryNameSpace::int2_*,int*,int);

CudaMatrix.cu文件的情况似乎与main函数是对称的。然而，编译器抱怨：

Error   19  error : Unresolved extern function '_ZN16LibraryNameSpace5int2_aSEi'    C:\Users\Documents\Project\Test\Testing_Files\ptxas simpleTest

请考虑：

1. 在将实现移到单独的文件之前，当在main文件中包含声明和实现时，一切都正常工作。
2. 有问题的指令是data_[i] = ob[i]。

有人知道这是怎么回事吗？

Answer 1

我在上述职位上所遵循的程序有两个问题：

1. ComplexTypes.cpp文件名必须将转换为ComplexTypes.cu，以便nvcc能够拦截CUDA关键字__device__和__host__。Talonmies在他的评论中指出了这一点。实际上，在发布之前，我已经将文件名从.cpp更改为.cu，但是编译器正在抱怨并显示相同的错误。因此，我巧妙地后退了一步；
2. 在Visual 2010中，必须使用查看->属性页；配置属性-> CUDA C/C++ ->公共->生成可释放的设备代码->是(-rdc=true)。这是单独汇编所必需的。事实上，在NVIDIA CUDA编译器驱动器NVCC，据说：

CUDA的工作方式是将设备代码嵌入到主机对象中。在整个程序编译过程中，将可执行的设备代码嵌入到主机对象中。在单独编译中，我们将可重定位的设备代码嵌入到主机对象中，并运行设备链接器(nvlink)将所有设备代码链接到一起。然后，由主机链接器将nvlink的输出与所有主机对象链接在一起，形成最终的可执行文件。--relocatable-device-code={true，false}选项控制可重定位 vs可执行设备代码的生成，该选项可缩短为-rdc={true，false}。