cuda 函数前缀 host device global __noinline 和 forceinline 简介

这些函数前缀在官方的文档里被称为函数执行环境标识符Function execution space specifiers，也就是他指明了这段函数是在哪里被调用的。

global

***在GPU上执行*** ，但是需要 ***在CPU端调用***
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21

PS.计算能力3.5以上的引入了 CUDA Dynamic Parallelism机制可以才devices端调用核函数

```__global__```修饰的函数必须采用```void```返回值，并且需要在调用时制定 ***运行的参数*** (也就是<<<>>>里的block数和线程数)

同时```__global___```函数是异步的，这也代表着函数没被执行完就返回了控制权，所以测量核函数的时间需要同步操作才能获得准确的结果。

## __device__

```__device__```函数是在GPU端调用且在GPU端执行的的函数```__global___```和```__device__```不能同时使用

```__device__``` 在编译器中会被编译为内联函数

## __host__

```__host__```定义了主机端的函数，就是在主机端执行，在主机端调用的函数，也就是我们正常的c/c++的函数，如果不加任何的修饰符，默认就是__host__ 函数，这些函数只为主机端编译

```__global__``` 和```__host__```不能同时使用，这是显而易见的

但是```__device__``` 和```__host__```可以同时使用，这个函数会同时为主机端和设备端编译。代码中可以用```__CUDA_ARCH__```这个宏来区分主机和设备端以及不同架构的代码

host device func()
{
#if CUDA_ARCH >= 700
// Device code path for compute capability 7.x
#elif CUDA_ARCH >= 600
// Device code path for compute capability 6.x
#elif CUDA_ARCH >= 500
// Device code path for compute capability 5.x
#elif CUDA_ARCH >= 300
// Device code path for compute capability 3.x
#elif !defined(CUDA_ARCH)
// Host code path
#endif
}


## __noinline__ 和 __forceinline__

前面我们说到```__device__```会被编译器编译为内联函数

```__noinline__```和 ```__forceinline__```则指定编译器不要内联，和强制内联这些函数。

这两个符号不能同时使用。