cuda 入门(一)

• 环境配置：

安装完cuda之后，查看cuda编译器nvcc能否能够正常工作。

• 第一个Helloworld

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

#include <stdio.h>

__global__ void kernel(void) {

}

int main() {
        kernel <<<1, 1>>> (); 
        // <<<1,1>>>代表一个kernel的grid中只有1个block，每个block中有1个thread。
        printf("Hello world!\n");
        return 0;
}

<<< >>> 为内核函数的执行参数，用于说明内核函数中的线程数量，以及线程是如何组织的。原型为：kernel_name<<<block_in_grid, thread_in_block>>>(argument list);

将上面的代码保存为一个cu文件，比如hello_world.cu, 然后编译该文件

#编译
nvcc hello_world.cu -o hello_cuda
#运行
hello_cuda

这样看，cuda程序的helloworld还是挺简单的。

• cuda计算程序

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

#include <stdio.h>

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

int main()
{
        int c;
        int *dev_c; // 定义在设备端的接收数据的指针
        cudaError_t cudaStatus;
        //为输入参数和输出参数分配内存
        cudaStatus = cudaMalloc((void**)&dev_c, sizeof(int));
        if (cudaStatus != cudaSuccess) {
                printf("cudaMalloc is failed!\n");
        }
        add<<<1, 1 >>>(2, 7, dev_c);
        cudaStatus = cudaMemcpy(&c, dev_c, sizeof(int), cudaMemcpyDeviceToHost);
        if (cudaStatus != cudaSuccess) {
                printf(" cudaMemcpyDeviceToHost is failed!\n");
        }
        cudaFree(dev_c);
        printf("2+7=%d\n", c);
        return 0;
}

add是真正在GPU上run的函数，程序刚开始在GPU memory中分配dev_c，将a+b的计算结果保存在GPU memory中，然后再将dev_c中的值copy到HOST端的c中，这样便可以print出来了。看程序还是很好理解的。

特别要注意的是，不能直接print dev_c，给dev_c分配了GPU memory空间，程序结束时，也要释放对应的GPU memory.

• cuda查询设备能力

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <stdio.h>

int main(void)
{
    cudaDeviceProp prop;
    int count;
    cudaError_t cudaStatus;
    cudaStatus = cudaGetDeviceCount(&count);
    for (int i = 0; i < count; i++) {
        cudaStatus = cudaGetDeviceProperties(&prop, i);
        printf("---General information for device %d ---\n", i);
        printf("name:%s\n", prop.name);
        printf("Max threads per block:%d\n", prop.maxThreadsPerBlock);
        printf("Max thread dimensions: (%d, %d, %d)\n", prop.maxThreadsDim[0], prop.maxThreadsDim[1], prop.maxThreadsDim[2]);
        printf("Max grid dimensions: (%d, %d, %d)\n", prop.maxGridSize[0], prop.maxGridSize[1], prop.maxGridSize[2]);
    }
    return 0;
}

• cudaDeviceProp数据结构定义

cudaDeviceProp数据类型针对函式cudaGetDeviceProperties定义的，cudaGetDeviceProperties函数的功能是取得支持GPU计算装置的相关属性，比如支持CUDA版本号装置的名称、内存的大小、最大的thread数目、执行单元的频率等。如下所示：

struct cudaDeviceProp {
    char name[256]; // 识别设备的ASCII字符串（比如，"GeForce GTX 1808 Ti"）
    size_t totalGlobalMem; // 全局内存大小
    size_t sharedMemPerBlock; // 每个block内共享内存的大小
    int regsPerBlock; // 每个block 32位寄存器的个数
    int warpSize; // warp大小
    size_t memPitch; // 内存中允许的最大间距字节数
    int maxThreadsPerBlock; // 每个Block中最大的线程数是多少
    int maxThreadsDim[3]; // 一个块中每个维度的最大线程数
    int maxGridSize[3]; // 一个网格的每个维度的块数量
    size_t totalConstMem; // 可用恒定内存量
    int major; // 该设备计算能力的主要修订版号
    int minor; // 设备计算能力的小修订版本号
    int clockRate; // 时钟速率
    size_t textureAlignment; // 该设备对纹理对齐的要求
    int deviceOverlap; // 一个布尔值，表示该装置是否能够同时进行cudamemcpy()和内核执行
    int multiProcessorCount; // 设备上的处理器的数量
    int kernelExecTimeoutEnabled; // 一个布尔值，该值表示在该设备上执行的内核是否有运行时的限制
    int integrated; // 返回一个布尔值，表示设备是否是一个集成的GPU（即部分的芯片组、没有独立显卡等）
    int canMapHostMemory; // 表示设备是否可以映射到CUDA设备主机内存地址空间的布尔值
    int computeMode; // 一个值，该值表示该设备的计算模式：默认值，专有的，或禁止的
    int maxTexture1D; // 一维纹理内存最大值
    int maxTexture2D[2]; // 二维纹理内存最大值
    int maxTexture3D[3]; // 三维纹理内存最大值
    int maxTexture2DArray[3]; // 二维纹理阵列支持的最大尺寸
    int concurrentKernels; // 一个布尔值，该值表示该设备是否支持在同一上下文中同时执行多个内核
｝