__assume() 似乎在 NVCC 中导致内部错误：“调用的参数数量错误......”答案

【问题标题】：__assume() seems to cause an internal error in NVCC: "Call has wrong number of parameters..."__assume() 似乎在 NVCC 中导致内部错误：“调用的参数数量错误......”
【发布时间】：2017-01-26 08:33:07
【问题描述】：

我有以下代码在较大代码库的上下文中导致错误：

struct Quadruple
{
    double _sum;
    double _corr;

    // Fast multiply by -1, 0 or +1 (doesn't need extended precision operations).
    __device__ inline Quadruple& MulSign(const int8_t sign);
};

// ...

__device__ inline Quadruple& Quadruple::MulSign(const int8_t sign)
{
    __assume(-1 <= sign && sign <= 1); // enable more optimization premises
    _sum *= sign;
    _corr *= sign;
    return *this;
}

// ...

int8_t y = /* ... read from memory ... */;
Quadruple a = /* ... computed here ... */;
Quadruple b = Quadruple(a).MulSign(y);

使用 CUDA 8RC + MSVC++2013 编译代码时出现以下错误：

1>  Compiling CUDA source file Example.gpu.cu...
1>  
1>  D:\ExDir>"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v8.0\bin\nvcc.exe" -gencode=arch=compute_20,code=\"sm_21,compute_20\" --use-local-env --cl-version 2013 -ccbin "C:\Program Files (x86)\Microsoft Visual Studio 12.0\VC\bin\x86_amd64"  -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v8.0\include" -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v8.0\include"     --keep-dir x64\Release -maxrregcount=0  --machine 64 --compile -cudart static     -D_DENSE_REP -DWIN32 -DWIN64 -DNDEBUG -D_CONSOLE -D_MBCS -Xcompiler "/EHsc /W4 /nologo /Ox /Zi  /MD " -o x64\Release\Example.gpu.cu.obj "D:\ExDir\Example.gpu.cu" 
1>ptxas C : /Users/ExUser/AppData/Local/Temp/tmpxft_00002ffc_00000000-4_CuSvm.gpu.ptx, line 2513; error : Call has wrong number of parameters
1>  ptxas fatal   : Ptx assembly aborted due to errors
1>  Example.gpu.cu
1>C:\Program Files (x86)\MSBuild\Microsoft.Cpp\v4.0\V120\BuildCustomizations\CUDA 8.0.targets(599,9): error MSB3721: The command ""C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v8.0\bin\nvcc.exe" -gencode=arch=compute_20,code=\"sm_21,compute_20\" --use-local-env --cl-version 2013 -ccbin "C:\Program Files (x86)\Microsoft Visual Studio 12.0\VC\bin\x86_amd64"  -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v8.0\include" -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v8.0\include"     --keep-dir x64\Release -maxrregcount=0  --machine 64 --compile -cudart static     -D_DENSE_REP -DWIN32 -DWIN64 -DNDEBUG -D_CONSOLE -D_MBCS -Xcompiler "/EHsc /W4 /nologo /Ox /Zi  /MD " -o x64\Release\Example.gpu.cu.obj "D:\ExDir\Example.gpu.cu"" exited with code 255.

注释掉__assume(-1 <= sign && sign <= 1);后，代码编译ok。

这是 NVCC 中的错误，还是 NVCC 不支持 MSVC++ 的 __assume，还是我做错了什么？是否有解决方法（当然，保留__assume 或等效项）？

【问题讨论】：

标签： visual-c++ optimization cuda intrinsics nvcc

【解决方案1】：

__assume() 是 Microsoft（以及随后的英特尔 icc）编译器内在函数。它不是 C++ 语言的一部分，并且在 CUDA（或在 GCC 或 Clang 中）不支持。

【讨论】：