CUDA实战笔记(2)——核函数传递参数

本系列代码托管于：https://github.com/chintsan-code/cuda_by_example
本篇使用的项目为：simple_kernel_params

// simple_kernel_params

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>
#include "../../common/book.h"

__global__ void add(int a, int b, int* c) {
    *c = a + b;
}

int main() {
    int c;
    int* dev_c;
    // 分配显存
    HANDLE_ERROR(cudaMalloc((void**)&dev_c, sizeof(int)));
    // 像调用C语言中的函数一样调用核函数
    add <<<1, 1 >>> (2, 7, dev_c);
    // 主机不能直接对dev_c所指的显存做操作，应该复制回主机内存
    HANDLE_ERROR(cudaMemcpy(&c, dev_c, sizeof(int), cudaMemcpyDeviceToHost));
    
    printf("2 + 7 = %d\n", c);
    // 最后要释放之前分配的显存
    cudaFree(dev_c);
    
    return 0;
}

当需要设备执行任何有用的操作时，都需要先分配内存。使用函数cudaMalloc来进行显存的分配，类似C语言的Malloc函数。

cudaMalloc函数的第一个参数是是一个指针，它指向新分配的显存地址。第二个参数是分配显存的大小。返回类型是void*，这个和C语言的Malloc有所区别。

HANDLE_ERROR是原书中用来处理异常的辅助宏(异常处理函数)，不用管。实际工程可以使用更加严谨的做法。

主机代码可以将设备指针dev_c作为参数传递给设备代码，但是不能直接使用这个指针进行读写，应该先使用cudaMemcpy()将设备内存中的数据复制到主机内存，才可以使用。cudaMemcpyDeviceToHost表示源指针指向设备内存，目标指针指向主机内存，如果是cudaMemcpyHostToDevice则相反。

总结：

• 可以将cudaMalloc()分配的指针传递给在设备上执行的代码
• 可以在设备代码中使用cudaMalloc()分配的指针进行内存读/写操作
• 可以将cudaMalloc()分配的指针传递给在主机上执行的函数
• 不能在主机代码中使用cudaMalloc()分配的指针进行内存读/写操作，要先使用cudaMemcpy()进行复制方可使用
• 主机只能访问主机内存，设备只能访问设备内存
• 释放由cudaMalloc()分配的内存，应该使用cudaFree()

参考：

• 《GPU高性能编程 CUDA实战》