c++ - CUDA 代码的 C++ 设计

Question

我有一段 C++ CUDA 代码，我必须编写它以在浮点中声明数据变量。我还必须重写以双精度声明数据变量的代码。

在 CUDA 中处理这种情况的好设计是什么？

我不希望有两组相同的代码，因为将来要进行任何更改，我将不得不更改两组其他相同的代码。我还想保持代码干净，而不需要在代码#ifdef中的 float 和 double 之间进行太多更改。

任何人都可以提出任何好的（在维护和“易于阅读”方面）设计吗？

Answer 1

CUDA 支持类型模板，毫无疑问，在您需要在同一代码中处理多种类型的情况下，它是实现内核代码的最有效方式。

作为一个简单的例子，考虑一个简单的 BLAS AXPY 类型的内核：

template<typename Real>
__global__ void axpy(const Real *x, Real *y, const int n, const Real a)
{
    int tid = threadIdx.x + blockIdx.x * blockDim.x;
    int stride = blockDim.x * gridDim.x;

    for(; tid<n; tid += stride) {
        Real yval = y[tid];
        yval += a * x[tid];
        y[tid] = yval;
    }
}

这个模板化的内核可以在不失一般性的情况下被实例化为双精度和单精度：

template axpy<float>(const float *, float *, const int, const float);
template axpy<double>(const double *, double *, const int, const double);

与所有最新版本的 CUDA 工具包一起提供的推力模板库广泛使用此工具来实现与类型无关的算法。

Answer 2

除了模板之外，您还可以使用单个 typedef 来实现您想要的：

typedef float mysize;  // or double

然后只需mysize在您将使用的地方使用floator double。

您可能对simpleTemplates 示例代码感兴趣，并且还有其他模板化的 CUDA 示例，除了推力，正如 talonmies 所述，它被广泛使用。Thrust 还为 C++ 程序员提供了许多其他好处。