Почему мы не можем разделить реализации __host__ и __device__?

Question

Если у нас есть __host__ __device__ Функция в CUDA, мы можем использовать макросы для выбора различных путей кода для кода на стороне хоста и на стороне устройства в его реализациях, например, так:

__host__ __device__ int foo(int x)
{
#ifdef CUDA_ARCH
    return x * 2;
#else
    return x;
#endif
}

но почему мы не можем написать:

__host__ __device__ int foo(int x);

__device__ int foo(int x) { return x * 2; }
__host__   int foo(int x) { return x; }

вместо?

Answer 1

Реализация CUDA C++ в Clang фактически поддерживает перегрузку __host__ а также __device__ потому что он считает квалификаторы пространства выполнения частью сигнатуры функции. Обратите внимание, однако, что даже там вы должны будете объявить две функции отдельно:

__device__ int foo(int x);
__host__ int foo(int x);

__device__ int foo(int x) { return x * 2; }
__host__   int foo(int x) { return x; }

проверить это здесь

Лично я не уверен, насколько желательным / важным, что на самом деле должно быть, хотя. Учтите, что вы можете просто определить foo(int x) в коде хоста за пределами вашего источника CUDA. Если кто-то сказал мне, что им нужно иметь разные реализации одной и той же функции для хоста и устройства, где по какой-либо причине необходимо определить версию хоста как часть источника CUDA, мое первоначальное интуитивное чувство будет заключаться в том, что, вероятно, что-то происходит немного странного направления. Если версия хоста делает что-то другое, не должно ли оно иметь другое имя? Если он логически делает то же самое, просто не используя графический процессор, то почему он должен быть частью источника CUDA? Я бы вообще выступал за то, чтобы сохранить как можно более четкое и строгое разделение между кодом хоста и устройства и сохранить любой код хоста внутри исходного кода CUDA до минимума. Даже если вы не заботитесь о чистоте своего кода, это по крайней мере сведет к минимуму шансы получить травму от всей магии компилятора, которая происходит под капотом…