Совместное использование констант между процессором и графическим процессором в CUDA ⇐ C++

[Anonymous]

Я хотел бы поделиться некоторыми постоянными между процессором и графическим процессором, чтобы разрешить выполнение того же кода (обернутый между ними). Это включает в себя некоторые параметры времени компиляции, которые наиболее разумно написаны в виде массивов, и я бы хотел, чтобы они были contexpr, так что компилятор может (в идеале) выдвигать массивы во время компиляции.

Код: Выделить всё

#include 
using fe_ftype = double;
__device__  constexpr fe_ftype vars[2] = {100.0, 300.0};
//__constant__ constexpr fe_ftype vars[2] = {100.0, 300.0};
const fe_ftype l = 3.0;

__global__ void foo() {
printf("%lf %lf %lf\n", vars[0], vars[1], l);
}

int main(void) {
foo();
cudaDeviceSynchronize();
printf("%lf %lf %lf\n", vars[0], vars[1], l);
}

(см. https://godbolt.org/z/19byj34f8)
Похоже, что в этом компиляции успешно, и я получаю одинаковые выводы как с графического процессора, так и с CPU, с константами, как константы, либо для CPU, либо для CPU, либо в CPU, или в REGERSTE. Тем не менее, компилятор жалуется с предупреждением #20091-D: переменная __device__ «vars» не может быть непосредственно прочитать в функции хоста . Если бы это не было постоянным, я бы получил доступ к какой -то случайной ненициализированной памяти на хосте, но с аннотацией коннидппр, кажется, успешно приводит ее на стороне хоста. Похоже, это также работает идентично, если я использую __constant __ вместо __device __ .
Могу ли я полагаться на это поведение? Разве предупреждение обычно не применяется к переменным contexpr?
дополнительное условие: я могу использовать C ++ 17, так что без постоянного. < /P>

Подробнее здесь: https://stackoverflow.com/questions/796 ... pu-in-cuda