Joh*_*ica 5 c++ cuda bit-manipulation
我有以下代码将一点变成一个字节.
__device__ UINT64 bitToByte(const UINT8 input) {
UINT64 b = ((0x8040201008040201ULL * input) >> 7) & 0x0101010101010101ULL;
//reverse the byte order <<-- this step is missing
return b;
}
Run Code Online (Sandbox Code Playgroud)
但是字节顺序错误,字节顺序相反.在CPU上我可以简单地bswap reg,reg解决这个问题,但是我在GPU上做了什么?
或者,我可以使用什么类似的技巧,以便正确地放置字节,即最高有效位转到最高有效字节,这样我就不需要bswap技巧.
为了反转字节顺序,可以使用相同的技巧来完成位提取,但是通过交换在乘法中执行移位的系数来完成。然而,为了避免乘法中的冲突,对于偶数位和奇数位必须分两步完成。这样,就有 2 个字节可以自由保存每次乘法的结果,这足以确保结果的完整性。
__device__ UINT64 bitToByte(const UINT8 input) {
UINT64 b = ( ((0x0002000800200080ULL * input) >> 7) & 0x0001000100010001ULL)
| ( ((0x0100040010004000ULL * input) >> 7) & 0x0100010001000100ULL);
return b;
}
Run Code Online (Sandbox Code Playgroud)
正如评论中所指出的,为了优化,可以分解这些变化。
__device__ UINT64 bitToByte(const UINT8 input) {
UINT64 b = ( ((0x0002000800200080ULL * input) & 0x0080008000800080ULL)
| ((0x0100040010004000ULL * input) & 0x8000800080008000ULL) )
>> 7 ;
return b;
}
Run Code Online (Sandbox Code Playgroud)