float(0.0) 与 CUDA 中的 0.0f

float(0.0) vs 0.0f in CUDA

我想编写对浮点精度和双精度都有效的代码。 我正在做这样的事情:

typedef real float;
//typedef real double;

__global__ void foo(real a, real *b){
  b[0] = real(0.5)*a;
}

int main(){
  real a = 1.0f;
  real *b;
  cudaMalloc(&f, sizeof(real));
  foo<<<1,1>>>(a,b);
  return 0;
}

这让我开始思考,我不想在做双精度时将常量的精度降低为 0.5f,但我不想在做单精度时将 0.5 提升为双精度!

所以我最终像示例中那样使用了运算符 real()。 在单精度模式下,如果我使用 real(0.5) 反汇编函数 'foo',我会发现没有提升到 double,而不是仅使用 0.5,那里会发生提升。

您可以使用以下方法检查:

$nvcc test.cu -arch=sm_52 -lineinfo --source-in-ptx -g -G -O0 ; cuobjdump -sass  a.out | grep "foo" -A 35

我明白了

/*0078*/                   FMUL R0, R0, 0.5;               /* 0x3868004000070000 */

使用实数(0.5) 或 0.5f 时 并且:

/*0078*/                   F2F.F64.F32 R4, R0;           /* 0x5ca8000000070b04 */
/*0088*/                   DMUL R4, R4, 0.5;               /* 0x3880004000070404 */
/*0090*/                   F2F.F32.F64 R0, R4;           /* 0x5ca8000000470e00 */

只写 0.5 时。

这听起来太明显了。但是因为我不知道 "real(0.5)" 在做什么,所以我不知道这是否只是编译器在这个非常特殊的情况下一起玩。反汇编代码在 real(0.5) 和 0.5f 中看起来是一样的!

所以问题依然存在:

real(0.5)(又名 float(0.5))到底在做什么?

float(0.5)和0.5f有区别吗? (OR double(0.5) 和 0.5)

我想这也适用于 C/C++。

real(0.5)(又名 float(0.5))到底在做什么?

real(0.5) function-style cast,在这种情况下它降低到 static_cast

real(0.5)
static_cast<real>(0.5) //exactly the same thing

这意味着 a 乘以一个 real 变量(在本例中为 float),这意味着不需要执行提升到 double,因为double * float 乘法就是这种情况。

float(0.5) 和 0.5f 有区别吗? (OR double(0.5) 和 0.5)

有人可能会争辩说 float0.5 的初始化可能会在运行时发生,但这对于任何现代编译器来说都是不现实的。它应该是一个空操作,它已经用于 OP。

除此之外,使用float(0.5f)与仅使用0.5f没有任何区别,double(0.5)0.5也是如此。