c++ - makefile: cpp + cu 文件 - undefined reference 错误

Question

考虑三个文件:main.cpp、func_prototypes.h 和 test_kernels.cu。我想从这三个文件中构建一个项目。我试图基于 CUDA 示例来构建“makefile”，但我失败了。 make 的执行返回 undefined reference 的错误。下面是我的三个文件和makefile

main.cpp

#include <iostream>

#include <cstdlib>
#include <stdlib.h>
#include <stdio.h>

#include <new>

#include <cuda.h>

using namespace std;

#include "func_prototypes.h"

typedef float mytype;

int main(){

    mytype *vec;
    unsigned int N = 1024;

    vec = new mytype[N];

    for(int i = 0; i < N; i++){
        vec[i] = i;
    }

    cout << "Calling CUDA function.\n";

    getSquares(vec,N);

    cout << "result:\n";

    for(int i = 0; i < N; i++){
        cout << vec[i] << " ";
    }

    ResetCUDA();
    return EXIT_SUCCESS;

}

func_prototypes.h

template <class type>
void getSquares(type *v, unsigned const int N);
void ResetCUDA();

test_kernels.cu

#include <cuda.h>
#include <new>

#define BlockSize 256

template <class type>
__global__
void getSquareKernel(type *v, unsigned const int N){
    int tIdx = blockIdx.x*blockDim.x + threadIdx.x;

    if(tIdx < N){
        v[tIdx] *= v[tIdx];
    }
}



template <class type>
void getSquares(type *v, unsigned const int N){

    int threads = BlockSize;
    int blocks = ceil(N/threads);

    type *d_v;
    cudaMalloc(&d_v,N*sizeof(type));
    cudaMemcpy(d_v,v,N*sizeof(type),cudaMemcpyHostToDevice);

    getSquareKernel<<<blocks,threads>>>(d_v,N);

    cudaMemcpy(v,d_v,N*sizeof(type),cudaMemcpyDeviceToHost);

    cudaFree(d_v);

}

void ResetCUDA(){
    cudaDeviceReset();
}

生成文件

############################# Makefile ##########################
CUDA_PATH       ?= /usr/local/cuda-5.0
CUDA_INC_PATH   ?= $(CUDA_PATH)/include
CUDA_BIN_PATH   ?= $(CUDA_PATH)/bin

ifeq ($(OS_SIZE),32)
    CUDA_LIB_PATH  ?= $(CUDA_PATH)/lib
else
    CUDA_LIB_PATH  ?= $(CUDA_PATH)/lib64
endif

ifeq ($(OS_SIZE),32)
    LDFLAGS :=  -L$(CUDA_LIB_PATH) -lcudart
    CPPFLAGS    :=  -m32
else
    LDFLAGS :=  -L$(CUDA_LIB_PATH) -lcudart
    CPPFLAGS    :=  -m64
endif
# Debug build flags
ifeq ($(dbg),1)
      CPPFLAGS      += -g
      NVCCFLAGS = -g -G
endif

INCLUDES    := -I$(CUDA_INC_PATH) -I. -I.. -I../../common/inc

CPP     =       icpc
NVCC        =       $(CUDA_BIN_PATH)/nvcc

SOURCE  =       main.cpp
AUX     =       test_kernels.cu

all:    test
test_kernels.o: $(AUX)
    $(NVCC) $(NVCCFLAGS) -o test_kernels.o -c $(AUX) $(NVCCFLAGS) $(INCLUDES)
main.o: $(SOURCE)
    $(CPP) $(CPPFLAGS) -o main.o -c $(SOURCE) $(CPPFLAGS) $(INCLUDES)
test:   test_kernels.o  main.o
    $(CPP) -o test test_kernels.o main.o $(LDFLAGS)
run: test
    ./test
clean:
    rm -rf test *.o

返回的错误是main.o:main.cpp:function main: error: undefined reference to 'void getSquares<float>(float*, unsigned int)'
make: *** [test] Error 1

有谁知道我的错误在哪里？

编辑:作为记录，我的操作系统是 Ubuntu 12.04 x86_64，内核 3.2.0-39

Answer 1

这是 standard gotcha 的体现在模板编译中。

您的宿主函数 getSquares 和内核 getSquareKernel 从未在定义它们的编译单元中实例化(即在 test_kernels.cu 中)。因此，编译器永远不会为 getSquares 发出任何代码，并且链接失败。

因为您在 test_kernels.cu 中使用组合的主机代码/设备代码编译轨迹，正确的解决方案是通过添加如下内容显式实例化您在 test_kernels.cu 中需要的模板代码的所有变体:

template __global__ void getSquareKernel<float>(float *, unsigned int);
template void getSquares<float>(float *, unsigned int);

到 test_kernels.cu 的底部。这将确保链接时需要链接的设备和主机代码实例都存在。

另一种选择是将包含 main 的文件更改为 .cu 文件，并将 test_kernels.cu 包含到该文件中，然后使用 nvcc 编译整个文件。在这种情况下，main() 中宿主类的实例化应该会触发同一编译单元中完整模板链的编译。

免责声明:我面前没有一台我可以对此进行测试的机器，所以至少要注意代码...