如何说服CMake使用CUDA fmax函数代替std cmath函数?

问题描述

说我具有以下功能

__global__ void testFunction(double *a,double *b) {
   unsigned int index = blockIdx.x * blockDim.x + threadIdx.x;
   a[index] = fmax(b[index],0.0f);
}

然后建筑物给出错误

错误:不允许从__global__函数(“ testFunction”)调用constexpr __host__函数(“ fmax”)。实验性标记“ --expt-relaxed-constexpr”可用于允许此操作。

apparently意味着我不能使用在标准库cmath中定义的函数fmax。但是,它也被定义为a CUDA device function,这是我首先要使用的那个。

所以问题是:如何告诉CMake(或任何编译器,如果可以的话)使用fmax的CUDA __device__定义而不是std cmath版本?

注意:我 不是 在此代码中使用using namespace std;和/或#include < cmath>

CMakeLists.txt:

cmake_minimum_required(VERSION 3.17)
project(NAME CUDA)

set(CMAKE_CUDA_STANDARD 14)

add_executable(NAME main.cu /*some other files*/)

set_target_properties(
        NAME
        PROPERTIES
        CUDA_SEParaBLE_COMPILATION ON)

解决方法

问题与CMake或链接无关。

CUDA使用模板重载在设备代码中提供本机数学库功能。根据文档,提供了fmax的两个版本(herehere)。他们是

float fmax(float,float)
double fmax(double double)

您的代码正在请求

double fmax(double,float)

因为0.0f是单个精度常数。对此没有任何本机重载,因此它属于CUDA工具链前端,并且编译器最终得出结论,您需要宿主函数,因此出错。

正确的代码应该是

a[index] = fmax(b[index],0.);

这将使用正确的双精度版本。