为什么我们不能拆分 __host__ 和 __device__ 实现？答案

【问题标题】：Why can't we split __host__ and __device__ implementations?为什么我们不能拆分 __host__ 和 __device__ 实现？
【发布时间】：2018-10-14 21:49:06
【问题描述】：

如果我们在 CUDA 中有一个__host__ __device__ 函数，我们可以使用宏在其实现中为主机端和设备端代码选择不同的代码路径，如下所示：

__host__ __device__ int foo(int x)
{
#ifdef CUDA_ARCH
    return x * 2;
#else
    return x;
#endif
}

但是为什么我们不能写：

__host__ __device__ int foo(int x);

__device__ int foo(int x) { return x * 2; }
__host__   int foo(int x) { return x; }

改为？

【问题讨论】：

唯一的答案是因为。
@talonmies：所以你说的是“提交功能请求”。
不，我是说这是 12 年前开发语言时有人故意选择的，事实就是这样。

标签： cuda nvcc

【解决方案1】：

CUDA C++ 的 Clang 实现实际上支持 __host__ 和 __device__ 因为它认为执行空间限定符是函数签名的一部分。但是请注意，即使在那里，您也必须分别声明这两个函数：

__device__ int foo(int x);
__host__ int foo(int x);

__device__ int foo(int x) { return x * 2; }
__host__   int foo(int x) { return x; }

test it out here

就我个人而言，我不确定这到底有多可取/重要。考虑到您可以在 CUDA 源代码之外的主机代码中定义 foo(int x)。如果有人告诉我他们需要为主机和设备提供相同功能的不同实现，其中由于某种原因需要将主机版本定义为 CUDA 源的一部分，我最初的直觉是可能会有一些事情发生一个奇怪的方向。如果主机版本做了不同的事情，它不应该有不同的名称吗？如果它在逻辑上做同样的事情只是不使用 GPU，那么为什么它必须是 CUDA 源的一部分？我通常主张在主机和设备代码之间保持尽可能干净和严格的分离，并将 CUDA 源代码中的任何主机代码保持在最低限度。即使您不关心代码的清洁度，这样做至少可以最大程度地减少被引擎盖下的所有编译器魔法伤害的机会……

【讨论】：

你不是在反对使用__host__ __device__ 函数吗？无论如何，我相信你的直觉是错误的。不是你在主机和设备上做别的事情，而是你用不同的方式来做。例如，位操作的实现可能不同。或者 - 某个功能可能不需要在设备上完成任何事情，而是在主机上完成某事。