CUDA 模板内核包装器

Question

我开始编写模拟并决定尝试使用更面向对象的方法。因此，我还决定在 CUDA 内核中使用模板参数，该参数指示模拟的空间维度。问题是，由于在头文件中实现模板函数的限制，我不得不使用一种复杂的方法来保持内核包装器可以从 .cpp 源文件中调用。

我的方法是重载 2 维和 3 维的包装函数。然后我有一个用于处理初始化和管理内核资源的包装类的类。不幸的是，由于我提到的限制，我必须为模板类保留两个成员，即

struct kernelWrapper{ KernelWrapper(Simulation&lt2&gt *simulation): d_(2), simulation2d_(simulation) {} KernelWrapper(Simulation&lt3&gt *simulation): d_(3), simulation3d_(simulation) {} process(void){ //wrapper function for kernel launching switch(d_){ case 2: kernel&lt2&gt&lt&lt&lt..., ...&gt&gt&gt(...); break; } case 3: kernel&lt3&gt&lt&lt&lt..., ...&gt&gt&gt(...); break; } default: break; } int d_; union{ Simulation&lt2&gt *simulation2d_; Simulation&lt3&gt *simulation3d_; }; union{ Lattice&lt2&gt *lattice2d_d; Lattice&lt3&gt *lattice3d_d; }; };

因此，我想知道您是否知道实现我正在尝试做的事情的更好方法，即为模板 CUDA 内核制作包装器。

更新：我想再添加一个我在发布上述帖子后发现的解决方案。正如C++ faq（第 13-15 点）所指出的，可以将模板实现放在源文件中，并显式实例化所需的模板，即在我的情况下是 2 维和 3 维。使用 C++11，可以更进一步，在模板定义中引入关键字extern 以节省一些编译/链接时间，还解释了here。

Answer 1

问题是，由于实现模板的限制 头文件中的函数，我不得不使用复杂的方法来 保持使内核包装器可从 .cpp 源文件调用。

在.cpp中编写模板声明代码是合法的

无论kernelWrapper 是在 .hpp 还是 .cpp 中，您都应该有一个看起来像这样的代码

template<int d_>
struct kernelWrapper
{
    KernelWrapper(Simulation<d_> *simulation) : simulation_(simulation)
    {}

    process(void)
    {
            kernel<d_><<<..., ...>>>(...);
    }

    Simulation<d_>* simulation_;
    Lattice<d_>*    lattice2d_;
};

同时避免使用 switch/case 来选择内核，使用类似：

int const max_dimension = 4;

template<int static_dimension>
void select_kernel(int dynamic_dimension)
{
    if(dynamic_dimension == static_dimension)
    {
        call_kernel<static_dimension>();
    }
    select_kernel<static_dimension+1>(dynamic_dimension);
}

template<>
void select_kernel<max_dimension>(int dynamic_dimension)
{
    // error message
}

void select_kernel(int dynamic_dimension)
{
    select_kernel<1>(dynamic_dimension);
}

如果这种选择很频繁，那么不使用模板是有意义的。