CUDA并行编程的一个例子

xmubingo

CUDA ，MPI，Hadoop都是并行运算的工具。
CUDA是基于GPU芯片计算。
简单这么理解。GPU有很多个核(几百个)，每个核可以跑一个线程，多个线程组成一个单位叫做块。

我们可以举个例子：
有三个向量 int a[10], b[10], c[10];
我们要计算a和b的向量之和存放到c中。
一般C语言：

1. for(int i=0; i<10; i++)
   
2.   c[i] = a[i] + b[i];

复制代码

CUDA编程做法：
GPU中的每个线程(核)有一个独立序号叫index，那么只要序号从0到9的线程执行c[index] = a[index] + b[index];就可以实现以上的for循环。
GPU的可贵之处就是，可以并发运行多个线程，相当于一个时间内赋值10次。

1. 
   
2. ////////////////////////
   
3. cuda.cu
   
4. ////////////////////////
   
5. #include <stdio.h>
   
6. #include <cuda_runtime.h>
   
7. 
   
8. /* 运行在GPU端的程序 */
   
9. __global__ void vectorADD(int* a, int* b, int* c)
   
10. {
    
11.      int index = threadIdx.x;//获得当前线程的序号
    
12.      if(index < blockDim.x)
    
13.          c[index] = a[index] + b[index];
    
14. }
    
15. 
    
16. int main ()
    
17. {
    
18.         /* 定义10个GPU运算线程 */
    
19.         int N = 10;
    
20.         
    
21.         /* 本地开辟三个数组存放我们要计算的内容 */
    
22.         int* h_a = (int*) malloc (N * sizeof(int));
    
23.         int* h_b = (int*) malloc (N * sizeof(int));
    
24.         int* h_c = (int*) malloc (N * sizeof(int));
    
25.         /* 初始化数组A, B和C */
    
26.         for(int i=0; i<N; i++)
    
27.         {
    
28.                 h_a[i] = i;
    
29.                 h_b[i] = i;
    
30.                 h_c[i] = 0;
    
31.         }
    
32. 
    
33.         /* 计算10个int型需要的空间 */
    
34.         int size = N * sizeof(int);
    
35.         
    
36.         /* 在GPU上分配同样大小的三个数组 */
    
37.         int* d_a;
    
38.         int* d_b;
    
39.         int* d_c;
    
40.         cudaMalloc((void**)&d_a, size);
    
41.         cudaMalloc((void**)&d_b, size);
    
42.         cudaMalloc((void**)&d_c, size);
    
43. 
    
44.         /* 把本地的数组拷贝进GPU内存 */
    
45.         cudaMemcpy(d_a, h_a, size, cudaMemcpyHostToDevice);
    
46.         cudaMemcpy(d_b, h_b, size, cudaMemcpyHostToDevice);
    
47.         cudaMemcpy(d_c, h_c, size, cudaMemcpyHostToDevice);
    
48.         
    
49.         /* 定义一个GPU运算块 由 10个运算线程组成 */
    
50.         dim3 DimBlock = N;
    
51.         /* 通知GPU用10个线程执行函数vectorADD */
    
52.         vectorADD<<<1, DimBlock>>>(d_a, d_b, d_c);
    
53.         /* 将GPU运算完的结果复制回本地 */
    
54.         cudaMemcpy(h_c, d_c, size, cudaMemcpyDeviceToHost);
    
55.         
    
56.         /* 释放GPU的内存 */
    
57.         cudaFree(d_a);
    
58.         cudaFree(d_b);
    
59.         cudaFree(d_c);
    
60. 
    
61.         /* 验证计算结果 */
    
62.         for(int j=0; j<N; j++)
    
63.                 printf("%d ", h_c[j]);
    
64.         printf("\n");
    
65. }
    

复制代码

cuda程序以cu为后缀，使用nvcc编译器。用法和gcc有相似之处。

描述有限，希望有兴趣的同学阅读http://datamining.xmu.edu.cn/bbs ... =467&extra=page%3D1 提供的文档。

我们实验室有三块GPU显卡，以及学院的CUDA服务器。设备齐全。有空之余大家多多交流。

zouquan

xmubingo 发表于 2012-3-2 10:06
我先写综述。时间太紧。
这里有四篇参考文章，应该不难。
http://blog.csdn.net/xhz1234/article/detai ...

好，我布置cwc来完成

xmubingo

zouquan 发表于 2012-3-1 21:37
可以实现的话就实现一下，陈伟程编了一个hadoop版的，和他的性能比较一下。

这个工作对聚类、生物信息 ...

我先写综述。时间太紧。
这里有四篇参考文章，应该不难。
http://blog.csdn.net/xhz1234/article/details/6222010
http://www.cnblogs.com/Fancyboy2 ... /04/28/1445636.html
http://topic.csdn.net/u/20090728 ... 9-2878dd5426d7.html
http://space.itpub.net/23057064/viewspace-676605

zouquan

xmubingo 发表于 2012-3-1 20:31
/* 运行在GPU端的程序 */
__global__ void vectorADD(int* a, int* b, int* c)
{

可以实现的话就实现一下，陈伟程编了一个hadoop版的，和他的性能比较一下。

这个工作对聚类、生物信息中的进化树构建、基因表达芯片数据分析都有用

xmubingo

zouquan 发表于 2012-3-1 19:51
在你上面的代码，哪里是gpu端的？

给定一个矩阵（double[][]），求每两行之间的距离，能不能用jcuda实 ...

/* 运行在GPU端的程序 */
__global__ void vectorADD(int* a, int* b, int* c)
{
     int index = threadIdx.x;//获得当前线程的序号
     if(index < blockDim.x)
         c[index] = a[index] + b[index];
}
这些就是gpu端的代码。

可以用Jcuda实现。

zouquan

xmubingo 发表于 2012-3-1 10:29
Jcuda的Host代码是可以用java写的。但是在gpu端的device代码还是需要用c写。

http://www.jcuda.org/tu ...

在你上面的代码，哪里是gpu端的？

给定一个矩阵（double[][]），求每两行之间的距离，能不能用jcuda实现？

xmubingo

zouquan 发表于 2012-2-29 23:18
very goooooooooooood，奖励50元！记账

另外，C太tmd难了，能不能弄个jCUDA的，让我学习一下

Jcuda的Host代码是可以用java写的。但是在gpu端的device代码还是需要用c写。

http://www.jcuda.org/tutorial/TutorialIndex.html

如果需要的话，我可以写个文档。Jcuda的思想是利用java的JNI组件来调用外部的C代码。因此cuda的代码编译仍然需要手动的用nvcc编译。然后才用java程序调用运行。

zouquan

very goooooooooooood，奖励50元！记账

另外，C太tmd难了，能不能弄个jCUDA的，让我学习一下