c++ - 如何仅针对一项功能禁用 Cuda 主机设备警告？

Question

我的 Cuda 代码中有一个 C++14 模板，该模板通过 lambda 闭包进行模板化，并且是 __host__ 和 __device__ 并且我收到警告:

warning: calling a __host__ function("Iter<(bool)1> ::Iter [subobject]")
         from a __host__ __device__ function("Horizontal::Horizontal")
         is not allowed

但是这是一个误报，因为它只是调用 __host__ 函数的模板的 __host__ 实例化，所以我希望仅为此抑制此警告一个模板定义。

我可以在模板之前添加此内容:

#pragma hd_warning_disable

警告消失了，但是，我担心我只希望对这个模板函数抑制它，而不是编译单元的其余部分。我无法轻易地将模板函数移动到编译单元的末尾。

我想要某种推送和弹出功能，但我在任何地方都找不到。

有没有办法在定义模板函数后使用编译指示重新启用高清警告？

我尝试过:

#pragma hd_warning_enable

但这不起作用:

test.cu:44:0: warning: ignoring #pragma 
hd_warning_enable  [-Wunknown-pragmas]
 #pragma hd_warning_enable

这是演示该问题的简单测试用例:

//#pragma hd_warning_disable

template<typename Lambda>
__host__ __device__
int hostDeviceFunction(const Lambda lambda)
{
    return lambda();
}


__device__
int deviceFunction()
{
    auto lambda = []() { return 0.0; };

    return hostDeviceFunction( lambda );
}

__host__
int hostFunction()
{
    auto lambda = []() { return 1.0; };

    return hostDeviceFunction( lambda );
}

这给出了这些警告:

test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed
test.cu(7): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction(Lambda) [with Lambda=lambda []()->double]" 
(24): here
test.cu(7): warning: calling a __host__ function(" const") from a __host__ __device__ function("hostDeviceFunction< ::> ") is not allowed

Answer 1

不需要像 #pragma hd_warning_enable 这样的东西如#pragma hd_warning_disable只影响它前面的函数。看起来这在任何文档中都找不到，但下面的示例表明了这种行为。

旁注:还有#pragma nv_exec_check_disable流行的图书馆已经迁移到该编译指示。参见例如this conversation关于它。

#include <iostream>
#include <cassert>

#pragma hd_warning_disable
//#pragma nv_exec_check_disable
template<typename Lambda>
__host__ __device__
int hostDeviceFunction1(const Lambda lambda)
{
    return lambda()*1.0;
}

__host__            
int hostFunction1()
{
    auto lambda = []() { return 1.0; };  
    return hostDeviceFunction1( lambda );
}

template<typename Lambda>
__host__ __device__
int hostDeviceFunction2(const Lambda lambda)
{                                       
    return lambda()*2.0;
}

__host__
int hostFunction2()
{
    auto lambda = []() { return 2.0; };  
    return hostDeviceFunction2( lambda );
}

int main()           
{ 
  std::cout << "hostFunction1: " << hostFunction1() << std::endl;
  assert(hostFunction1() == 1.0);

  std::cout << "hostFunction2: " << hostFunction2() << std::endl;
  assert(hostFunction2() == 4.0);

  return 0;
}

$ nvcc pragma_test.cu 
pragma_test.cu(24): warning: calling a __host__ function from a __host__ __device__ function is not allowed
          detected during instantiation of "int hostDeviceFunction2(Lambda) [with Lambda=lambda []()->double]" 
(31): here