#include <stdio.h>
#include <stdlib.h>
#include <cuda.h>
#include <time.h>
#define N 512
__global__ void vec_block_add(int *a, int *b, int *c)
{
c[blockIdx.x] = a[blockIdx.x] + b[blockIdx.x];
}
void rand_ints(int *arr, int count)
{
srand(time(NULL));
for(int i=0; i<count; i++)
{
arr[i] = rand()%100;
}
}
int main()
{
int *a, *b, *c;
int *d_a, *d_b, *d_c;
int size = N*sizeof(int);
clock_t start, finish;
cudaMalloc((void**)&d_a, size);
cudaMalloc((void**)&d_b, size);
cudaMalloc((void**)&d_c, size);
a=(int*)malloc(size);
rand_ints(a, N);
b=(int*)malloc(size);
rand_ints(b, N);
c=(int*)malloc(size);
rand_ints(c, N);
cudaMemcpy(d_a, a, size, cudaMemcpyHostToDevice);
cudaMemcpy(d_b, b, size, cudaMemcpyHostToDevice);
start = clock();
vec_block_add<<<100, 1>>>(d_a, d_b, d_c);
finish = clock();
double time = (finish - start)/(double)CLOCKS_PER_SEC;
cudaMemcpy(c, d_c, size, cudaMemcpyDeviceToHost);
for(int i=0; i<N; i++)
{
printf("%-5d: a:%-5d b:%-5d c:%-5d\n",i,a[i],b[i],c[i]);
}
printf("%lf\n",time);
cudaFree(d_a);
cudaFree(d_b);
cudaFree(d_c);
free(a);
free(b);
free(c);
return 0;
}
程序说明
程序有两个关键部分:
c[blockIdx.x] = a[blockIdx.x] + b[blockIdx.x];
vec_block_add<<<100, 1>>>(d_a, d_b, d_c);
add<<<N,1>>>
N表示同时调用N次add函数,这是实现并行的向量相加的语句。每个被并行调用的add函数称之为一个块(block)。
块的集合称之为网格(grid).
每个块可以使用索引值blockIdx.x
函数并行执行的过程,相当于由GPU这个加速器使用硬件的方式进行了循环展开,而blockIdx.x就是用来定位当前执行的是循环的哪个部分。通过使用blockIdx.x作为索引,每个块可以处理数组元素中的一部分。从硬件的角度看,相当于同时有多个块在并行执行:
块0: c[0]=a[0]+b[0]
块1: c[1]=a[1]+b[1]
块2: c[2]=a[2]+b[2]
块3: c[3]=a[3]+b[3]
….
程序中通过时间函数比较了并行计算的块数目不同时的计算时间。
本文参考自hipercome博文,向作者致谢!