Cuda函数调用

一. 函数声明

#include <iostream> 

_global_ void kernel(void){ } 

int main（ void）{ 
    kernel<<<1, 1>>>(); 
    printf("Hello, World！\ n");
    return 0; 
}

__global_修饰符提示该函数由主机（Host）发起对设备（Device）的调用。函数执行将在gpu设备上，而非在cpu上。默认使用gpu的全局内存。

<<< >>的表示运行线程参数，1, 1表示用gpu的1个线程块，该块中包含1个线程。具体将稍后说明。

二. 函数调用顺序

1. 分配内存

cudaMalloc((void**)&dev_c, sizeof(int));

该函数将在gpu设备上申请分配内存空间：

&dev_c: 主机指针，指向设备新分配的内存首地址.

sizeof(int): 设备分配的内存大小，假设分配的数据类型是1个int.

2. 内存拷贝

cudaMemcpy(&c, dev_c, sizeof(int), cudaMemcpyDeviceToHost)

该函数将设备指针c所指内容拷贝至主机指针dev_c:

&c: 设备指针，c是指针所指内容，通常为__global__里面的返回计算结果

dev_c：主机指针

cudaMemcpyDeviceToHost: 枚举类型，表示拷贝source和target.这里指的是从设备copy至主机。同理还有:

        cudaMemcpyDeviceToDevice: 两个gpu设备之间相互拷贝,

        cudaMemcpyHo stToDevice:  数据从主机拷贝至设备.

3. 释放内存

cudaFree(dev_c)

不再使用gpu时，应当删除(释放)指向设备的内容指针。

4. 获取当前支持cuda的GPU数量

int count = 0;
cudaGetDeviceCount(&count)

函数调用后，count值即为当前gpu数量.

5.   获取当前cuda设备属性

cudaDeviceProp prop；
cudaGetDeviceProperties(&prop, i):    //单个gpu, i = 0

当有多个gpu时，需要由i来控制gpu的index。函数调用后可以std::cout查看prop。

三. 线程块（Block）

1. kernel<<<N, M>>>：表示有N个线程块，每个线程块里面包含M个线程。N， M既可以是一维，也可以是二维。

        （1）若N，M是一维， blockDim.x = M， blockDim.y = 1 

                偏移量offset_x = threadIdx.x + blockIdx.x * blockDim.x;

                偏移量offset_y = 0;

                总偏移量offset = offset_x;

           (2)  若N是二维， M是一维： dim3 N(dim1 dim1, 1);    blockDim.x = M，blockDim.y = 1 

                偏移量offset_x = threadIdx.x + blockIdx.x * blockDim.x;

                偏移量offset_y = blockIdx.y * blockDim.y;

                总偏移量offset = offset_x + offset_y * blockIdx.x;

             (3) 若N是二维，M是二维：dim3 N(dim1, dim1, 1); dim3 M(dim2, dim2, 1).

                        blockDim.x = blockDim.y = dim2.

                偏移量offset_x = threadIdx.x + blockIdx.x * blockDim.x;

                偏移量offset_y = threadIdy.y + blockIdx.y * blockDim.y;

                总偏移量offset = offset_x + offset_y * blockIdx.x * gridDim.x；

将线程块中细分多维(grid)可以解决线程块数量的硬件限制, 否则最多只能有65536个线程）