2013-08-21 2 views
0
__device__ __inline__ double ld_gbl_cg(const double *addr) { 
      double return_value; 
      asm("ld.global.cg.f64 %0, [%1];" : "=d"(return_value) : "l"(addr)); 
      return return_value; 
     } 

Приведенный выше код отсюда: CUDA disable L1 cache only for one variableпримитивный тип данных в PTX

По мнению автора, «д» означает с плавающей точкой, «г» означает Int.

Я хочу написать небольшую часть встроенного кода asm, я хочу знать, что символ для остальной части примитивных переменных типа (например, unsigned short, unsigned long long, float-32 и т. Д.), Я не могу найти, что от ptx isa.

Я использую букву «l» для представления unsigned long long, это правильно?

ответ

2

Вы можете найти их here, но для полноты картины, буквы соответствуют нижележащих PTX типов регистров:

"h" = .u16 reg 
"r" = .u32 reg 
"l" = .u64 reg 
"f" = .f32 reg 
"d" = .f64 reg 

так что unsigned long long карты на «л» (для 64 битного целого PTX регистра).

+0

Спасибо за это, и int64 использует тот же символ/тип reg, что и uint64 здесь? – user2188453

+1

Письмо с ограничениями совпадает между регистром 's64' и регистром' u64', но фактический тип регистра отличается. Если вы изучите ссылку, которую @talonmies дал, вы сразу заметите, что после вышеприведенной таблицы есть пример, который передает параметр в регистр 's64', используя букву ограничения' l'. Аналогичным образом вы увидите примеры, в которых используется буква ограничения 'r' для передачи значений в/из' s32' регистров. –

Смежные вопросы