CUDA 和链接器错误答案

【问题标题】：CUDA and Linker errorsCUDA 和链接器错误
【发布时间】：2011-07-14 01:14:41
【问题描述】：

这可能是与Linker errors 2005 and 1169 (multiply defined symbols) when using CUDA __device__ functions (should be inline by default) 类似的问题，但不完全一样。尝试在 VS2010 上构建项目（使用已显示在其他地方工作的代码）时，我遇到了几个 LNK2005 错误。我无计可施。

例如，我有以下三个文件：transposeGPU.h、transposeGPU.cu 和 transposeCUDA.cu。 transposeGPU.h可以总结如下：

void transposeGPU(float *d_dst, size_t dst_pitch,
    float *d_src, size_t src_pitch,
    unsigned int width, unsigned int height);

即，没有任何包含的单个声明。该函数的定义见transposeGPU.cu，总结如下：

#include <stdio.h>
#include "../transposeGPU.h"
#include "../helper_funcs.h"

#include "transposeCUDA.cu"

void
transposeGPU(float *d_dst, size_t dst_pitch,
    float *d_src, size_t src_pitch,
    unsigned int width, unsigned int height)
{
    // execution configuration parameters
    dim3 threads(16, 16);
    dim3 grid(iDivUp(width, 16), iDivUp(height, 16));
    size_t shared_mem_size =
        (threads.x * threads.y + (threads.y - 1)) * sizeof(float);

    transposeCUDA<<<grid, threads, shared_mem_size>>>(
        d_dst, dst_pitch / sizeof(float),
        d_src, src_pitch / sizeof(float),
        width, height);
}

即tranposeGPU.cu包括它的头文件和transposeCUDA.cu，除了定义transposeGPU()和调用transposeCUDA()，后者在transposeCUDA.cu中找到。现在，transposeCUDA.cu 按预期定义了函数：

#include "common_kernel.h"

__global__ void
transposeCUDA(
    float *g_dst, size_t s_dst_pitch,
    const float *g_src, size_t s_src_pitch,
    unsigned int img_width, unsigned int img_height)
{
// several lines of code...
}

看起来一切正常，但我仍然在transposeGPU.obj 中得到error LNK2005: "void __cdecl __device_stub__Z13transposeCUDAPfjPKfjjj(float *,unsigned int,float const *,unsigned int,unsigned int,unsigned int)" (?__device_stub__Z13transposeCUDAPfjPKfjjj@@YAXPAMIPBMIII@Z) already defined in transposeCUDA.obj。

这和其他大约 20 个类似的链接器错误。为什么？没有明显的重新定义发生。任何帮助将不胜感激。

【问题讨论】：

标签： c++ linker cuda

【解决方案1】：

如果同时编译 transposeCUDA.cu 和 transposeGPU.cu，则会发生重新定义，因为定义出现在两个翻译单元中。您不应该 #include transposeCUDA.cu 并将 nvcc 应用于该文件。

【讨论】：

我不确定我是否理解。我应该为transposeCUDA.cu 编写一个头文件并将其包含在transposeGPU.cu 中吗？ transposeGPU.cu 至少需要声明 transposeCUDA() 才能工作。
最简单的解决方案是在 transposeGPU 中只提供一个声明

【解决方案2】：

澄清一下：__device__ 函数是内联的（至少在 Fermi 之前），但 __global__ 不是——毕竟，您不能将 GPU 代码内联到 CPU 可执行函数中。全局函数可以获取它们的地址，唯一的区别是地址指向 GPU 内存（类似于存储在 GPU 上的数据的普通指针看起来只是普通指针）。

正如 William Pursell 所说，如果你编译你的全局函数两次，你会得到两个具有相同定义的函数，从而导致链接器错误。

【讨论】：