CUDA的学习

前面几天写了三维重建中的特征提取部分，下面接着写，不过今天写一下CUDA的内容，这个下面要用到，要学习，首先装环境，装了CUDA5.0，网上有一个windows7+CUDA5.0的教程，挺好的，按照那个基本上没有问题，不过实际在运行的过程中发现了一个问题，就是不同通过windows的远程桌面连接，远程桌面不能调用显卡。

于是想远程操作的则可以下载一个Symantec pcAnywhere，远程控制软件。好了不多说，我们就先看第一个程序。也就是CUDA5.0自带的demo。

把整个程序列上来，然后加了点注释，这个也只是参考加上本人的理解，也不知道正确与否。望大家指正。

#include "cuda_runtime.h" //使用 runtime API

#include "device_launch_parameters.h"

#include <stdio.h>

cudaError_t addWithCuda(int *c, const int *a, const int *b, size_t size);

// 显示芯片上执行的程序。在 CUDA 中，在函式前面加上

// __global__ 表示这个函式是要在显示芯片上执行的。

// 在显示芯片上执行的程序有一些限制，例如它不能有传回值。

// 接下来是要让 CUDA 执行这个函式。

__global__ void addKernel(int *c, const int *a, const int *b)

{

// threadIdx  是 CUDA 的一个内部的变量，表示目前的 thread 是

//第几个 thread（由 0 开始计算）这边会有 size =5 个 threads

//每一个的 threadIdx.x  则分别会是 0 ~ 5。利用这个变量，我们

//就可以让每一份函式执行时，对整个数据不同的部分进行相加。

int i = threadIdx.x;

c[i] = a[i] + b[i];

}

int main()

{

const int arraySize = 5;

const int a[arraySize] = { 1, 2, 3, 4, 5 };

const int b[arraySize] = { 10, 20, 30, 40, 50 };

int c[arraySize] = { 0 };

// Add vectors in parallel.

cudaError_t cudaStatus = addWithCuda(c, a, b, arraySize);

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "addWithCuda failed!");

return 1;

}

printf("{1,2,3,4,5} + {10,20,30,40,50} = {%d,%d,%d,%d,%d}\n",

c[0], c[1], c[2], c[3], c[4]);

// cudaDeviceReset must be called before exiting in order for profiling and

// tracing tools such as Nsight and Visual Profiler to show complete traces.

cudaStatus = cudaDeviceReset();

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "cudaDeviceReset failed!");

return 1;

}

return 0;

}

// Helper function for using CUDA to add vectors in parallel.

cudaError_t addWithCuda(int *c, const int *a, const int *b, size_t size)

{

int *dev_a = 0;

int *dev_b = 0;

int *dev_c = 0;

cudaError_t cudaStatus;

// Choose which GPU to run on, change this on a multi-GPU system.

// 选择哪个GPU去运行，这个在多GPU系统上修改。

cudaStatus = cudaSetDevice(0);

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "cudaSetDevice failed!  Do you have a CUDA-capable GPU installed?");

goto Error;

}

// Allocate GPU buffers for three vectors (two input, one output)

// 给三个向量（两个输入，一个输出）分配GPU缓存.

// 要利用 CUDA 进行计算之前，要先把数据复制到显卡内存中，// 才能让显示芯片使用。因此，

// 需要取得一块适当大小的显卡内存，再把产生好的数据复制进去

cudaStatus = cudaMalloc((void**)&dev_c, size * sizeof(int));

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "cudaMalloc failed!");

goto Error;

}

cudaStatus = cudaMalloc((void**)&dev_a, size * sizeof(int));

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "cudaMalloc failed!");

goto Error;

}

cudaStatus = cudaMalloc((void**)&dev_b, size * sizeof(int));

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "cudaMalloc failed!");

goto Error;

}

// Copy input vectors from host memory to GPU buffers.

// 复制输入向量从主存到显卡内存。

// 通过调用cudaMalloc 取得一块显卡内存，然后通过过

// cudaMemcpy 将上述两个数组复制到显卡内存中.

// cudaMalloc 和 cudaMemcpy 的用法和一般的 malloc 及

// memcpy 类似，不过 cudaMemcpy 则多出一个参数，指示复制内// 存的方向。在这里因为是从主内存复制到显卡内存，所以使用

// cudaMemcpyHostToDevice。如果是从显卡内存到主内存，则使用 // cudaMemcpyDeviceToHost。

cudaStatus = cudaMemcpy(dev_a, a, size * sizeof(int), cudaMemcpyHostToDevice);

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "cudaMemcpy failed!");

goto Error;

    }

cudaStatus = cudaMemcpy(dev_b, b, size * sizeof(int), cudaMemcpyHostToDevice);

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "cudaMemcpy failed!");

goto Error;

}

// Launch a kernel on the GPU with one thread for each element.

// 函式名称<<<block 数目, thread 数目, shared memory 大小

// >>>(参数...);

//在 CUDA 中，thread 是可以分组的，也就是 block。一个 block 中//的 thread，具有一个共享的 shared memory，也可以进行同步工作。//不同 block 之间的 thread 则不行。在我们的程序中，其实不太需//要进行 thread 的同步动作，因此我们可以使用多个 block 来进一//步增加 thread 的数目。就是把这个1改变

addKernel<<<1, size>>>(dev_c, dev_a, dev_b);

// cudaDeviceSynchronize waits for the kernel to finish, and returns

// any errors encountered during the launch.

cudaStatus = cudaDeviceSynchronize();

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "cudaDeviceSynchronize returned error code %d after launching addKernel!\n", cudaStatus);

goto Error;

}

// Copy output vector from GPU buffer to host memory.

cudaStatus = cudaMemcpy(c, dev_c, size * sizeof(int), cudaMemcpyDeviceToHost);

if (cudaStatus != cudaSuccess) {

fprintf(stderr, "cudaMemcpy failed!");

goto Error;

}

Error:

cudaFree(dev_c);

cudaFree(dev_a);

cudaFree(dev_b);

return cudaStatus;

}

使用5.0的时候要注意跟以前版本的不一样的地方，一个是这个版本里面没有了

#include "cuda_util.h"

也就是以前使用的向CUDA_SAFE_CALL之类的就不好使用了

修改以前版本的可以去掉通过上面的方法使用cudaError_t，也可以使用

#include <helper_functions.h>#include <helper_cuda.h>

不过不是C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v5.0\include

下的头文件最好不要引用，以免跟上面一样。

在对于CUDA 的操作中，开始的时候要将主存上的数据复制到GPU上，而GPU上又分为好几种

几种的区别以及何时用，也就是一个优化的过程，这个是个难点。

下面就主要的是一个并行算法的问题，这个也是最关键的。

下面再把Blocks 和 Threads和grid的关系图放上来

这样一个基本的CUDA程序就基本上了解了，不过要理解的还有很多，先写这么多。