Idiom for CUDA class static member in device code?

Question

So, I've got a C++14 library that I'm porting to CUDA 9. I actually have (I think) a pretty good knowledge of CUDA, but I haven't done any direct work in it since CUDA 6.

Typically, I use a lot of templates and small classes in my code. It's surprised me that one still cannot have a static __device__ class member in CUDA 9, but global variables are fine. Is there a good idiom or workaround for this? What do people typically do?

Edit: I should be clear, I mean specifically for templated classes. If the class isn't templated, it's pretty straightforward.

Edit 2: Here is some example code

In normal host-side C++ I do this:

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

On the device, this won't compile, so what's a good equivalent?

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

Answer 1

Your original code works if you compile your CUDA with clang.