设备代码中CUDA类静态成员的成语？

Posted 2021-04-21

所以，我有一个C ++ 14库，我正在移植到CUDA 9.我实际上（我认为）对CUDA非常了解，但是自从CUDA 6以来我没有做过任何直接的工作。

通常，我在代码中使用了很多模板和小类。令我感到惊讶的是，在CUDA 9中仍然没有一个static __device__类成员，但全局变量很好。这有一个很好的习语或解决方法吗？人们通常做什么？

编辑：我应该清楚，我的意思是专门针对模板化课程。如果这个类没有模板化，那就非常简单了。

编辑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。

进一步的修改 - 基本上是一种风格问题 - 可以添加一种非技术静态但又有种类的吸气剂：

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，你的原始代码是有效的。