CUDA编程中global 和 device具体的使用方式和使用场景分析

原创于 2025-03-19 13:48:56 发布 · 1.1k 阅读

18 ·

CC 4.0 BY-SA版权

文章标签：

#开发语言 #人工智能 #算法

《深度学习》专栏收录该内容

12 篇文章

订阅专栏

在CUDA编程中，__global__ 和 __device__ 是两种不同的函数修饰符，它们分别用于定义可由主机端调用的 GPU 内核函数和只能在 GPU 设备端调用的辅助函数。了解这两者的具体使用方式和使用场景非常重要，以便优化代码并正确地设计主机和设备之间的交互。

1. `global` 关键字

定义：

__global__ 修饰符用于定义 GPU 内核函数（Kernel），即那些由主机代码发起并在 GPU 上并行执行的函数。
__global__ 函数必须返回 void 类型，且通常没有返回值。
主机端代码通过特定的语法 <<<...>>> 来启动 __global__ 函数。

使用方式：

__global__ 函数可以通过主机端调用（CPU 调用），并且 GPU 会并行地执行该函数。
函数调用时，需要指定 线程块数 和 每个线程块中的线程数，即通过 <<<gridDim, blockDim>>> 语法来设置。
__global__ 函数默认在 GPU 上执行，并且每个线程在执行时可以访问其线程 ID 及其线程块 ID，便于在并行执行时确定每个线程的计算任务。

代码示例：

__global__ void add(int *a, int *b, int *c, int N) {
    int idx = blockIdx.x * blockDim.x + threadIdx.x;  // 计算全局线程索引
    if (idx < N) {
        c[idx] = a[idx] + b[idx];  // 执行加法操作
    }
}

int main() {
    int N = 1000;
    int *a, *b, *c;
    // 在主机端分配内存并初始化数据
    // 通过 cudaMalloc 分配设备内存，初始化数据等
    add<<<(N + 255) / 256, 256>>>(a, b, c, N);  // 启动内核函数
    // 结果处理
}

使用场景：

执行并行计算：当需要将计算任务分配到多个线程并行执行时使用 __global__。例如，数组加法、矩阵乘法、图像处理等。
从主机调用并发任务：当你从主机（CPU）端启动并发任务，并希望在 GPU 上进行计算时使用 __global__。

2. `device` 关键字

定义：

__device__ 修饰符用于定义在 GPU 设备端执行的函数，这些函数只能在设备代码中调用，不能直接由主机代码调用。
__device__ 函数允许在设备代码内部进行复用，例如从 __global__ 函数中调用或在设备代码中进行其他计算。

使用方式：

__device__ 函数只能由其他 __device__ 或 __global__ 函数调用，不能从主机代码直接调用。
这些函数通常用于封装某些操作，例如计算的细节部分，供多个内核函数共享。

代码示例：

__device__ int multiply(int x, int y) {
    return x * y;
}

__global__ void kernel(int *a, int *b, int *c, int N) {
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    if (idx < N) {
        c[idx] = multiply(a[idx], b[idx]);  // 调用 __device__ 函数
    }
}

使用场景：

函数复用：当你需要将一些计算操作封装成函数，以便在多个内核函数或设备端的多个地方复用时，可以使用 __device__ 函数。例如，执行一些基础运算、验证操作等。
计算细节封装：将一些复杂的计算操作从内核函数中提取出来，通过 __device__ 函数进行封装，使得内核函数本身更加简洁易读。

3. 区别与联系

特性	`__global__`	`__device__`
调用方式	由主机端调用，通过 `<<<...>>>` 语法启动	只能在设备端（GPU）被调用
执行位置	在设备（GPU）并行执行	在设备（GPU）上执行，但不能从主机端调用
返回值	必须返回 `void`	可以有返回值
调用场景	用于定义内核函数，从主机端启动并在 GPU 上并行执行	用于定义仅限设备调用的函数，通常用于在 GPU 上执行辅助操作
参数传递	由主机端传入参数，并执行并行计算	由设备端其他函数传递参数，并执行计算

实际使用场景示例：

__global__ 用于内核函数：
- 在计算密集型任务中，如矩阵乘法、图像处理、深度学习的卷积操作等。每个线程处理输入数据的一个部分，结果可以并行地计算。
__device__ 用于设备端共享函数：
- 例如，定义一个设备端函数用于计算一小部分复杂的数学操作，这样可以让多个内核共享同一段代码，从而避免重复编写相同的逻辑。