Cộng 2 vector – Lập trình song song trên GPU bằng CUDA

ngoton 08/09/2021 Parallel Programming on a GPU Leave a comment 763 Views

Mục lục

Bài viết đầu tiên trong chuỗi bài viết về Lập trình song song trên GPU bằng CUDA.

Chúng ta sẽ cài đặt thuật toán Cộng 2 vector trên CPU và GPU. Sau đó, chúng ta sẽ so sánh tốc độ thực thi trên một số cấu hình khác nhau.

Nội dung

Kích thước mỗi vector: 2^13 + 1

Phát sinh ngẫu nhiên giá trị của các vector đầu vào trong [0, 1]

So sánh thời gian chạy với các kích thước block khác nhau

Thực hiện

Thuật toán cộng 2 vector khá đơn giản

for (int i = 0; i < n; i++)
      out[i] = in1[i] + in2[i];

1 2	for (int i = 0; i < n; i++) out[i] = in1[i] + in2[i];

Chúng ta sẽ cho mỗi thread đảm nhiệm vị trí ứng với mỗi phần tử của vector.

int i = blockIdx.x * blockDim.x + threadIdx.x;
 
if (i < n)
     out[i] = in1[i] + in2[i];

int i = blockIdx.x * blockDim.x + threadIdx.x;

if (i < n)

out[i] = in1[i] + in2[i];

Nếu máy tính bạn nào không có GPU thì có thể lên Google Colab và cài plugin nvcc để chạy nhé.

!pip install git+git://github.com/andreinechaev/nvcc4jupyter.git
%load_ext nvcc_plugin

%%cu
#include <stdio.h>
#include <stdlib.h>
#include <math.h>
#include <time.h>

void addVecOnHost(float *in1, float *in2, float *out, int n)
{
    for (int i = 0; i < n; i++)
      out[i] = in1[i] + in2[i];
}

__global__ void addVecOnDevice(float *d_in1, float *d_in2, float *d_out, int n)
{
    int i = blockIdx.x * blockDim.x + threadIdx.x;
 
    if (i < n)
        d_out[i] = d_in1[i] + d_in2[i];
}

int main(int argc, char* argv[])
{
    int sizeArr[10] = {64, 256, 1024, 4069, 16384, 65536, 262144, 1048576, 4194304, 16777216};
  
    for(int i = 0; i < 10; i++)
    {
      int n = sizeArr[i];
      float *in1, *in2, *out;
      float *d_in1, *d_in2, *d_out;
      size_t bytes = n * sizeof(float);
  
      in1 = (float*)malloc(bytes);
      in2 = (float*)malloc(bytes);
      out = (float*)malloc(bytes);
  
      cudaMalloc(&d_in1, bytes);
      cudaMalloc(&d_in2, bytes);
      cudaMalloc(&d_out, bytes);
  
      for(int j = 0; j < n; j++ )
      {
          in1[j] = sin(j)*sin(j);
          in2[j] = cos(j)*cos(j);
      }
     
      clock_t begin, end;
     
      begin = clock();
      addVecOnHost(in1, in2, out, n);
      end = clock();

      double duration = double(end-begin)/CLOCKS_PER_SEC * 1000;
     
      float hostSum = 0;
      for(int l = 0; l < n; l++)
          hostSum += out[l];
      
      printf("(Vector size %d):\n", n);
      printf("CPU:\n");
      printf("\tResult: %.f\n", hostSum/n);
      printf("\tTime: %.3f ms \n", duration);
     
      cudaEvent_t start, stop;
      cudaEventCreate(&start);
      cudaEventCreate(&stop);
     
      cudaMemcpy(d_in1, in1, bytes, cudaMemcpyHostToDevice);
      cudaMemcpy(d_in2, in2, bytes, cudaMemcpyHostToDevice);
  
      dim3 blockSize(512);
      dim3 gridSize((n - 1) / blockSize.x + 1);
     
      cudaEventRecord(start);
      addVecOnDevice<<<gridSize, blockSize>>>(d_in1, d_in2, d_out, n);
      cudaEventRecord(stop);
  
      cudaMemcpy(out, d_out, bytes, cudaMemcpyDeviceToHost);
     
      cudaEventSynchronize(stop);
      float milliseconds = 0;
      cudaEventElapsedTime(&milliseconds, start, stop);
  
      float deviceSum = 0;
      for(int l = 0; l < n; l++)
          deviceSum += out[l];
     
      printf("GPU:\n");
      printf("\tResult: %.f\n", deviceSum/n);
      printf("\tTime: %.3f ms \n", milliseconds);
      printf("Host/Device: %.3f\n", duration/milliseconds);
      printf("___________________________________\n");
     
      cudaFree(d_in1);
      cudaFree(d_in2);
      cudaFree(d_out);
  
      free(in1);
      free(in2);
      free(out);
    }
 
    return 0;
}

100

101

102

103

104

105

106

107

108

!pip install git+git://github.com/andreinechaev/nvcc4jupyter.git

%load_ext nvcc_plugin

%%cu

#include <stdio.h>

#include <stdlib.h>

#include <math.h>

#include <time.h>

void addVecOnHost(float *in1, float *in2, float *out, int n)

{

for (int i = 0; i < n; i++)

out[i] = in1[i] + in2[i];

}

__global__ void addVecOnDevice(float *d_in1, float *d_in2, float *d_out, int n)

{

int i = blockIdx.x * blockDim.x + threadIdx.x;

if (i < n)

d_out[i] = d_in1[i] + d_in2[i];

}

int main(int argc, char* argv[])

{

int sizeArr[10] = {64, 256, 1024, 4069, 16384, 65536, 262144, 1048576, 4194304, 16777216};

for(int i = 0; i < 10; i++)

{

int n = sizeArr[i];

float *in1, *in2, *out;

float *d_in1, *d_in2, *d_out;

size_t bytes = n * sizeof(float);

in1 = (float*)malloc(bytes);

in2 = (float*)malloc(bytes);

out = (float*)malloc(bytes);

cudaMalloc(&d_in1, bytes);

cudaMalloc(&d_in2, bytes);

cudaMalloc(&d_out, bytes);

for(int j = 0; j < n; j++ )

{

in1[j] = sin(j)*sin(j);

in2[j] = cos(j)*cos(j);

}

clock_t begin, end;

begin = clock();

addVecOnHost(in1, in2, out, n);

end = clock();

double duration = double(end-begin)/CLOCKS_PER_SEC * 1000;

float hostSum = 0;

for(int l = 0; l < n; l++)

hostSum += out[l];

printf("(Vector size %d):\n", n);

printf("CPU:\n");

printf("\tResult: %.f\n", hostSum/n);

printf("\tTime: %.3f ms \n", duration);

cudaEvent_t start, stop;

cudaEventCreate(&start);

cudaEventCreate(&stop);

cudaMemcpy(d_in1, in1, bytes, cudaMemcpyHostToDevice);

cudaMemcpy(d_in2, in2, bytes, cudaMemcpyHostToDevice);

dim3 blockSize(512);

dim3 gridSize((n - 1) / blockSize.x + 1);

cudaEventRecord(start);

addVecOnDevice<<<gridSize, blockSize>>>(d_in1, d_in2, d_out, n);

cudaEventRecord(stop);

cudaMemcpy(out, d_out, bytes, cudaMemcpyDeviceToHost);

cudaEventSynchronize(stop);

float milliseconds = 0;

cudaEventElapsedTime(&milliseconds, start, stop);

float deviceSum = 0;

for(int l = 0; l < n; l++)

deviceSum += out[l];

printf("GPU:\n");

printf("\tResult: %.f\n", deviceSum/n);

printf("\tTime: %.3f ms \n", milliseconds);

printf("Host/Device: %.3f\n", duration/milliseconds);

printf("___________________________________\n");

cudaFree(d_in1);

cudaFree(d_in2);

cudaFree(d_out);

free(in1);

free(in2);

free(out);

}

return 0;

}

Kết quả

(Vector size 64):
CPU:
	Result: 1
	Time: 0.002 ms 
GPU:
	Result: 1
	Time: 0.016 ms 
Host/Device: 0.122
___________________________________
(Vector size 256):
CPU:
	Result: 1
	Time: 0.002 ms 
GPU:
	Result: 1
	Time: 0.011 ms 
Host/Device: 0.188
___________________________________
(Vector size 1024):
CPU:
	Result: 1
	Time: 0.004 ms 
GPU:
	Result: 1
	Time: 0.012 ms 
Host/Device: 0.346
___________________________________
(Vector size 4069):
CPU:
	Result: 1
	Time: 0.014 ms 
GPU:
	Result: 1
	Time: 0.011 ms 
Host/Device: 1.322
___________________________________
(Vector size 16384):
CPU:
	Result: 1
	Time: 0.052 ms 
GPU:
	Result: 1
	Time: 0.007 ms 
Host/Device: 7.888
___________________________________
(Vector size 65536):
CPU:
	Result: 1
	Time: 0.206 ms 
GPU:
	Result: 1
	Time: 0.011 ms 
Host/Device: 17.982
___________________________________
(Vector size 262144):
CPU:
	Result: 1
	Time: 0.852 ms 
GPU:
	Result: 1
	Time: 0.018 ms 
Host/Device: 47.124
___________________________________
(Vector size 1048576):
CPU:
	Result: 1
	Time: 4.833 ms 
GPU:
	Result: 1
	Time: 0.056 ms 
Host/Device: 86.058
___________________________________
(Vector size 4194304):
CPU:
	Result: 1
	Time: 18.357 ms 
GPU:
	Result: 1
	Time: 0.196 ms 
Host/Device: 93.827
___________________________________
(Vector size 16777216):
CPU:
	Result: 1
	Time: 73.773 ms 
GPU:
	Result: 1
	Time: 0.773 ms 
Host/Device: 95.438
___________________________________

(Vector size 64):

CPU:

Result: 1

Time: 0.002 ms

GPU:

Result: 1

Time: 0.016 ms

Host/Device: 0.122

___________________________________

(Vector size 256):

CPU:

Result: 1

Time: 0.002 ms

GPU:

Result: 1

Time: 0.011 ms

Host/Device: 0.188

___________________________________

(Vector size 1024):

CPU:

Result: 1

Time: 0.004 ms

GPU:

Result: 1

Time: 0.012 ms

Host/Device: 0.346

___________________________________

(Vector size 4069):

CPU:

Result: 1

Time: 0.014 ms

GPU:

Result: 1

Time: 0.011 ms

Host/Device: 1.322

___________________________________

(Vector size 16384):

CPU:

Result: 1

Time: 0.052 ms

GPU:

Result: 1

Time: 0.007 ms

Host/Device: 7.888

___________________________________

(Vector size 65536):

CPU:

Result: 1

Time: 0.206 ms

GPU:

Result: 1

Time: 0.011 ms

Host/Device: 17.982

___________________________________

(Vector size 262144):

CPU:

Result: 1

Time: 0.852 ms

GPU:

Result: 1

Time: 0.018 ms

Host/Device: 47.124

___________________________________

(Vector size 1048576):

CPU:

Result: 1

Time: 4.833 ms

GPU:

Result: 1

Time: 0.056 ms

Host/Device: 86.058

___________________________________

(Vector size 4194304):

CPU:

Result: 1

Time: 18.357 ms

GPU:

Result: 1

Time: 0.196 ms

Host/Device: 93.827

___________________________________

(Vector size 16777216):

CPU:

Result: 1

Time: 73.773 ms

GPU:

Result: 1

Time: 0.773 ms

Host/Device: 95.438

___________________________________

Programming, Network, Office, Hardware, Software, Design, Photography.. I am just a GEEK

Hi, I am

Ngô Tôn

I am a programmer.

Cộng 2 vector – Lập trình song song trên GPU bằng CUDA

Related Articles

Nội dung

Thực hiện

Kết quả

About ngoton

Check Also

Nhân 2 ma trận – Lập trình song song trên GPU bằng CUDA

Leave a Reply

Lập trình hướng đối tượng trong Golang – OOP in Go

Java Interview Questions And Answers

Xây dựng website tự động tạo văn bản sử dụng OpenAI GPT-2

Đếm số lượng chỗ trống trong bãi đỗ xe sử dụng OpenCV Python

Ứng dụng AI cảnh báo người lạ đột nhập

Tổng hợp 1000 bài tập lập trình C/C++ từ cơ bản tới nâng cao

Hàm làm tròn số ROUND, FLOOR, CEIL – Nhật ký lập trình [Chapter 7]

Xác định góc phần tư thứ mấy trên đường tròn lượng giác

Hàm swap

Hoán vị 2 số không dùng biến tạm

Lập trình hướng đối tượng trong Golang – OOP in Go

Java Interview Questions And Answers

Xây dựng website tự động tạo văn bản sử dụng OpenAI GPT-2

Đếm số lượng chỗ trống trong bãi đỗ xe sử dụng OpenCV Python

Ứng dụng AI cảnh báo người lạ đột nhập

Đếm số lượng chỗ trống trong bãi đỗ xe sử dụng OpenCV Python

Liệt kê các số nguyên tố nhỏ hơn N

Bài 3: Công cụ Move-Zoom và Hand – Photoshop CS6

Hướng dẫn xây dựng Chatbot bằng NLTK và Keras

In ra các phần tử chẵn và nhỏ hơn 20