Какие настоящие языковые конструкции C++ поддерживаются кодом устройства CUDA?

Приложение D версии 3.2 документации CUDA относится к поддержке C++ в коде устройства CUDA.
Четко указано, что CUDA поддерживает «классы для устройств с вычислительными возможностями 2.x». Однако я работаю с устройствами с вычислительными возможностями 1.1 и 1.3 и могу использовать эту функцию!

Например, этот код работает:

// class definition voluntary simplified
class Foo {
  private:
    int x_;

  public:
    __device__ Foo() { x_ = 42; }
    __device__ void bar() { return x_; }
};


//kernel using the previous class
__global__ void testKernel(uint32_t* ddata) {
    Foo f;
    ddata[threadIdx.x] = f.bar(); 
}

Я также могу использовать широко распространенные библиотеки, такие как Thrust::random классы случайной генерации. Мое единственное предположение состоит в том, что я могу сделать это благодаря автоматическому встраиванию помеченной __device__ функции, но это не объясняет также обработку переменных-членов.

Вы когда-нибудь использовали такие возможности в таких же условиях, или вы можете объяснить мне, почему мой код CUDA ведет себя именно так? Что-то не так в справочнике?


person jopasserat    schedule 04.02.2011    source источник


Ответы (2)


Официально CUDA не поддерживает классы на устройствах до версии 2.0.

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

Это означает, что вы можете использовать следующие конструкции C++:

  • Видимость (общедоступная/защищенная/частная)
  • невиртуальное наследование
  • все шаблонное программирование и метапрограммирование (пока не наткнетесь на баги nvcc; их довольно много начиная с версии 3.2)
  • конструкторы (кроме случаев, когда объект объявлен в __ разделяемой __ памяти)
  • пространства имен

Вы не можете использовать следующее:

  • новые и удаляющие операторы (я считаю, что устройства> = 2.0 могут это сделать)
  • виртуальные методы (требуются переходы по адресу переменной)
  • рекурсия функций (требуется вызов функций)
  • исключения

На самом деле, все примеры в главе D.6 Руководства по программированию CUDA могут компилироваться для устройств ‹2.0.

person CygnusX1    schedule 26.02.2011
comment
Ваш ответ содержит интересные моменты, касающиеся ограничений. Необходимость времени компиляции, по-видимому, является тем моментом, который следует учитывать при работе с устройствами с вычислительными возможностями ‹ 2.X. - person jopasserat; 01.03.2011
comment
Изучив проблему, я не могу (и, вероятно, не найду) лучшего ответа, чем ваш. Таким образом, я полагаюсь на ваш опыт :) В любом случае, теперь у меня есть C2050 \o/, поэтому я могу сравнить свои исполнения, чтобы убедиться, что мой код остается переносимым. Спасибо за ваш ответ @CygnusX1 - person jopasserat; 16.06.2011

Некоторые функциональные возможности класса C++ будут работать, однако в Руководстве по программированию в основном говорится, что они не полностью поддерживаются, и поэтому не все функциональные возможности класса C++ будут работать. Если вы можете сделать то, что вы хотите сделать, то вы должны идти вперед!

person Tom    schedule 05.02.2011
comment
Хорошо, но мне нужно точно знать, что сработает. Я намерен использовать эти функции в библиотеке, поэтому я не могу полагаться на них, если они небезопасны. Мы легко могли себе представить, что этот код может привести к хаосу при определенных обстоятельствах. Что мне действительно нужно, так это точная граница между тем, что я должен делать или не делать. - person jopasserat; 06.02.2011