c++ - CUDA 和模板 : specialization declaration needed?

Question

我有一个模板化包装函数，它调用在 .cu 文件中定义的内核 (__global__)

template<typename T, class M> 
__global__ void compute_kernel(T* input, T* output, n) {
    M m;
    // compute stuff using m
};

template<typename T, class M> 
void compute(T* input, T* output, int n) {
    // ... compute blocks, threads, etc.
    compute_kernel<T,M> <<<dim_grid, dim_block>>>(input, output, n);
    // ...
};

和一个头文件包含在只有声明的主机代码中

template<typename T, class M> 
void compute(T* input, T* output, int n);

但是，使用任意模板参数从主机调用 compute() 时，编译失败并显示 undefined reference to 'void reduce(...)' 并且仅当我将特化声明添加到 .cu 文件的末尾进行代码编译:

template void
compute<int, Method1<int> >(int* input, int* output, int n);

template void
compute<float, Method1<float> >(float* input, float* output, int n);

template void
compute<int, Method2<int> >(int* input, int* output, int n);

template void
compute<float, Method2<float> >(float* input, float* output, int n);

那么，是否有必要专门化每个模板化函数以使其可从主机调用？ (这是一个很大的缺点)

感谢您的评论!

Answer 1

这是一个 C++ FAQ ，不限于 CUDA。

如果您在 .cpp 或 .cu 文件中有模板实现，那么当您编译该翻译单元时，编译器不可能知道您需要什么样的模板参数排列。因此，当您链接时，您会收到错误。

您可以将实现放在头文件中(在这种情况下，您需要在 .cu 文件中实例化，因为它包含 CUDA)，或者您必须显式实例化所有必需的排列。如果您必须执行其中许多操作，那么您可以使用宏来实例化所有排列。