设备代码中 CUDA class 静态成员的成语?

Idiom for CUDA class static member in device code?

所以,我有一个要移植到 CUDA 9 的 C++14 库。我实际上(我认为)对 CUDA 有很好的了解,但我没有在这方面做过任何直接的工作它自 CUDA 6.

通常,我在代码中使用大量模板和小型 classes。令我惊讶的是,在 CUDA 9 中仍然不能有 static __device__ class 成员,但全局变量没问题。对此有好的成语或解决方法吗?人们通常做什么?

编辑:我应该清楚,我的意思是专门针对模板化的 classes。如果 class 没有模板化,那就很简单了。

编辑 2:这是一些示例代码

在正常的主机端 C++ 中,我这样做:

template <typename T>
class MyClass {
    static T my_static_member;
};

在设备上,这不会编译,那么什么是好的等效项?

template <typename T>
class MyClass {
    static __device__ T my_static_member;
};

您可以使用(模板化的)全局变量而不是静态成员,并且可能将它们放在详细命名空间中:

namespace detail {
namespace MyClass {

template <typename T> __device__ T my_static_member;

} // namespace detail
} // namespace MyClass

template <typename T>
class MyClass {

    // ... you use detail::MyClass::my_static_member<T> in the code
};

...但这不需要将 --std=C++14 传递给 nvcc

进一步修改 - 基本上是风格问题 - 可以添加 not-technically-static-but-kind-of getter:

template <typename T>
class MyClass {
    const T& my_static_member() const {
        return detail::MyClass::my_static_member<T>;
    }
    // ... you use my_static_member() in the code
};

如果您compile your CUDA with clang,您的原始代码就可以工作。