向量相加其三(CUDA+C语言实现)
·
测试机配置
硬件信息
CPU
厂家:Intel
型号:Intel®Pentium® CPU 5405U
核数:4
频率:2.3GHz 2.3GHz
指令集:不支持AVX2/AVX512
GPU1
厂家:Intel
型号:Intel®UHD Graphics 610
显存容量:120MB
显存频率:300MHz
显存带宽:17.1GB/s
GPU2
厂家:NVIDIA
型号:NVIDIA GeForce MX250
显存容量:2GB
显存频率:1519MHz
显存带宽:48.1GB/s
C语言源代码
#include <stdio.h>
#include <stdlib.h>
#include <sys/time.h>
#include <omp.h>
#include <immintrin.h>
__global__
void vecadd_cuda(int *d_c, int *d_a, int *d_b, int n)
{
const int tid = blockIdx.x * blockDim.x + threadIdx.x;
if (tid < n) d_c[tid] = d_a[tid] + d_b[tid];
}
int main(int argc, char **argv) {
int n = atoi(argv[1]);
int repeat = 10;
//create vectors
int *a = (int *)malloc(sizeof(int) * n);
int *b = (int *)malloc(sizeof(int) * n);
int *c = (int *)malloc(sizeof(int) * n);
for(int i = 0; i < n; i++) {
a[i] = 1;
b[i] = 2;
}
struct timeval t1, t2;
int *d_a, *d_b, *d_c;
cudaMalloc((void **)&d_a, n * sizeof(int));
cudaMalloc((void **)&d_b, n * sizeof(int));
cudaMalloc((void **)&d_c, n * sizeof(int));
cudaMemcpy(d_a, a, n * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(d_a, b, n * sizeof(int), cudaMemcpyHostToDevice);
//serial C
gettimeofday(&t1, NULL);
for (int i = 0; i < repeat; i++) {
for (int j = 0; j < n; j++) c[j] = a[j] + b[j];
}
gettimeofday(&t2, NULL);
double time_serial = (t2.tv_sec - t1.tv_sec) * 1000.0 + (t2.tv_usec - t1.tv_usec) / 1000.0;
time_serial /= repeat;
time_serial /= 1000.0;
printf("C serial takes %f sec\n", time_serial);
//OpenMP parallel version
gettimeofday(&t1, NULL);
for (int i = 0; i < repeat; i++) {
#pragma omp parallel for
for (int j = 0; j < n; j++) c[j] = a[j] + b[j];
}
gettimeofday(&t2, NULL);
time_serial = (t2.tv_sec - t1.tv_sec) * 1000.0 + (t2.tv_usec - t1.tv_usec) / 1000.0;
time_serial /= repeat;
time_serial /= 1000.0;
printf("C OpenMP takes %f sec\n", time_serial);
//OpenMP+AVX2 parallel version
gettimeofday(&t1, NULL);
for (int i = 0; i < repeat; i++) {
int loop = n/8;
#pragma omp parallel for
for (int j = 0; j < loop; j++)
{
__m256i aavx2 = _mm256_loadu_si256((__m256i*)(&a[j * 8]));
__m256i bavx2 = _mm256_loadu_si256((__m256i*)(&b[j * 8]));
__m256i cavx2 = _mm256_add_epi32(aavx2, bavx2);
_mm256_storeu_si256((__m256i*)(&c[j * 8]), cavx2);
}
}
gettimeofday(&t2, NULL);
time_serial = (t2.tv_sec - t1.tv_sec) * 1000.0 + (t2.tv_usec - t1.tv_usec) / 1000.0;
time_serial /= repeat;
time_serial /= 1000.0;
printf("C OpenMP+avx2 takes %f sec\n", time_serial);
//CUDA parallel version
gettimeofday(&t1, NULL);
for (int i = 0; i < repeat; i++)
{
int nthreads = 128;
int nblocks = ceil((double)n / (double)nthreads);
vecadd_cuda<<<nblocks, nthreads>>>(d_c, d_a, d_b, n);
}
cudaDeviceSynchronize();
gettimeofday(&t2, NULL);
time_serial = (t2.tv_sec - t1.tv_sec) * 1000.0 + (t2.tv_usec - t1.tv_usec) / 1000.0;
time_serial /= repeat;
time_serial /= 1000.0;
printf("CUDA takes %f sec\n", time_serial);
cudaFree(d_a);
cudaFree(d_b);
cudaFree(d_c);
free(a);
free(b);
free(c);
return 0;
}
运行结果及其分析
kendrick@kendrick-LAPTOP:~/Desktop/vecadd-c$ ./vecadd 10
C serial takes 0.0000000000 sec
CUDA takes 0.0000048000 sec
kendrick@kendrick-LAPTOP:~/Desktop/vecadd-c$ ./vecadd 100
C serial takes 0.0000001000 sec
CUDA takes 0.0000047000 sec
kendrick@kendrick-LAPTOP:~/Desktop/vecadd-c$ ./vecadd 1000
C serial takes 0.0000003000 sec
CUDA takes 0.0000047000 sec
kendrick@kendrick-LAPTOP:~/Desktop/vecadd-c$ ./vecadd 10000
C serial takes 0.0000047000 sec
CUDA takes 0.0000046000 sec
kendrick@kendrick-LAPTOP:~/Desktop/vecadd-c$ ./vecadd 100000
C serial takes 0.0000516000 sec
CUDA takes 0.0000321000 sec
kendrick@kendrick-LAPTOP:~/Desktop/vecadd-c$ ./vecadd 1000000
C serial takes 0.0011572000 sec
CUDA takes 0.0002842000 sec
kendrick@kendrick-LAPTOP:~/Desktop/vecadd-c$ ./vecadd 10000000
C serial takes 0.0123769000 sec
CUDA takes 0.0027254000 sec
kendrick@kendrick-LAPTOP:~/Desktop/vecadd-c$ ./vecadd 100000000
C serial takes 0.1254904000 sec
CUDA takes 0.0283268000 sec
由此可见cuda将数据放在GPU上计算将大大加速向量计算,只是将数据从CPU中传入GPU这一过程将耗费很多的事件,上面的运行结果中并未体现出来。
欢迎来到FlagOS开发社区,这里是一个汇聚了AI开发者、数据科学家、机器学习爱好者以及业界专家的活力平台。我们致力于成为业内领先的Triton技术交流与应用分享的殿堂,为推动人工智能技术的普及与深化应用贡献力量。
更多推荐



所有评论(0)