Aaaaa988

Шпора на ПГП

Jun 16th, 2020
150
Never
Not a member of Pastebin yet? Sign Up, it unlocks many cool features!
  1. #include <stdio.h>
  2. #include <cublas_v2.h>
  3. #include <thrust/host_vector.h>
  4. #include <thrust/device_vector.h>
  5. #include <thrust/transform.h>
  6. #include <thrust/fill.h>
  7. #include <thrust/sequence.h>
  8. #define N 9000000
  9.  
  10. //SAXPY — это комбинация скалярного умножения и сложения векторов, то есть: в качестве входных данных
  11. //принимается два 32-разрядных вектора X и Y с N элементами каждый и скалярное значение A. Далее происходит
  12. //умножение каждого элемента X [i] на A, а получение результата приходит к Y [i].
  13.  
  14. //Функтор - объект, который можно использовать как функцию
  15. //Функтор получающий на вход два аргумента и возвращающий результат их обработки как переопределенный оператор ().
  16. struct saxpyFunctor {//Бинарный оператор
  17.     const float a;
  18.     saxpyFunctor(float _a) : a(_a) {}
  19.     __host__ __device__ float operator()(float x, float y) {
  20.         return a * x + y;
  21.     }
  22. };
  23.  
  24. //threadIdx.x это потоки в блоке
  25. //blockIdx.x это сам блок
  26. //blockDim.x это размерность блока
  27. //представим что у нас есть цикл и потоки допустим на блок у нас от 1 до 5
  28. //блоки исчисляются с 0 и далее а размерность блока допустим 5
  29. //тогда первый цикл первый поток + 0 блок * 5 = 1 индекс в виюдюхе
  30. //второй поток + 0 блок *5 = 2 индекс
  31. //потом допустим блок мы этот забили таким образом, значит переменная blockIdx.x + 1 сделала
  32. //получаем 1 поток в блоке + 1 блок *5 = 6 индекс в видюхе
  33. __global__ void cudaSaxpy(float alpha, float* x, float* y) {
  34.     int i = threadIdx.x + blockIdx.x * blockDim.x;
  35.     //определяем индекс, по координате потока в блоке + координаты блока в сетке * размер блока
  36.     y[i] = alpha * x[i] + y[i];
  37.     //вычисляем Saxpy где S - float - тоесть вещественные числа, а axpy это a*x+y (ax plus y = axpy)
  38. }
  39.  
  40. //В случае бинарной операции на вход поступают два входных вектора, один выходной и сама операция.
  41. void saxpy(float a, thrust::device_vector<float>& x, thrust::device_vector<float>& y) {//функция для траст библиотеки
  42.     saxpyFunctor func(a);// инициализация бинарного оператора.
  43.     thrust::transform(x.begin(), x.end(), y.begin(), y.begin(), func);
  44.     //*происходит трансформация, от начала вектора х и У до их конца, соглсно функции func
  45.     //происходит проход от начала x и y, где согласно БИНАРНОЙ функции SAXPY мы выводим результат в операторный вектор y,
  46.     //тоесть мы берем из У информацию, обсчитываем вместе с Х и кладем обратно в ту же ячейку вектора У откуда брали только что данные
  47. }
  48.  
  49. void CudaKernel() {
  50.     float elapsedTime;// затраченное время
  51.     cudaEvent_t start, stop;// Объекты событий
  52.     float* x_d, * x_h, * y_h, * y_d;// вещественные указатели на x и y девайса и хоста
  53.     cudaEventCreate(&start); // функция создания события
  54.     cudaEventCreate(&stop);
  55.     //выделение глобальной памяти где x_d указатель на память в девайсе, N - размерность
  56.     cudaMalloc((void**)&x_d, N * sizeof(float));
  57.     cudaMalloc((void**)&y_d, N * sizeof(float));
  58.     x_h = (float*)calloc(N, sizeof(float));//выделение памяти на хосте
  59.     y_h = (float*)calloc(N, sizeof(float));
  60.  
  61.     for (int i = 0; i < N; i++) {
  62.         x_h[i] = i;
  63.         y_h[i] = 0.87;
  64.     }
  65.     // перенос вектора с хоста на девайс
  66.     cudaMemcpy(x_d, x_h, N * sizeof(float), cudaMemcpyHostToDevice);
  67.     cudaMemcpy(y_d, y_h, N * sizeof(float), cudaMemcpyHostToDevice);
  68.     // запускаем событие, для начала отсчета таймера
  69.     cudaEventRecord(start, 0);
  70.     /// <<< ... >>> это синтаксис конфигурации выполнения
  71.     //Вызов ядра <<< Число блоков, число нитей в блоке>> ( аргументы функции) туда мы передаем альфу, и векторы на девайсе
  72.     cudaSaxpy << < N / 256, 256 >> > (3.0, x_d, y_d);
  73.     cudaDeviceSynchronize();//Блокируется до тех пор, пока устройство не выполнит все предыдущие запрошенные задачи
  74.     cudaEventRecord(stop, 0);// фиксируем событие окончания вычислений
  75.     //Данная функция ожидает окончание работы всех нитей GPU и прохождение заданного event’а и только потом отдает управление вызывающей программе.
  76.     cudaEventSynchronize(stop);//Ожидание до завершения всех работ, захваченных в данный момент в события
  77.     cudaEventElapsedTime(&elapsedTime, start, stop);//Вычисляет прошедшее время между событиями.
  78.     cudaMemcpy(y_h, y_d, N * sizeof(float), cudaMemcpyDeviceToHost);//копирование с девайса на хост
  79.     printf("CUDA Kernel Time:\n \t %f ms\n", elapsedTime);
  80.     free(x_h);//высвобождение памяти на хосте
  81.     free(y_h);
  82.     cudaFree(y_d);// высвобождение глобальной памяти на девайсе
  83.     cudaFree(x_d);
  84.     cudaEventDestroy(start);
  85.     cudaEventDestroy(stop);
  86. }
  87.  
  88. void cudaCublas() {
  89.     float elapsedTime;
  90.     cudaEvent_t start, stop;
  91.     float* cx_d, * cx_h, * cy_h, * cy_d;
  92.     cudaEventCreate(&start);
  93.     cudaEventCreate(&stop);
  94.     //выделение закрепленной (pinned) памяти, особый участок памяти на ОЗУ
  95.     cudaMallocHost((void**)&cx_h, N * sizeof(float));
  96.     cudaMallocHost((void**)&cy_h, N * sizeof(float));
  97.     // выделение на девайсе глобальной памяти
  98.     cudaMalloc((void**)&cx_d, N * sizeof(float));
  99.     cudaMalloc((void**)&cy_d, N * sizeof(float));
  100.  
  101.     for (int i = 0; i < N; i++) {// заполнение на хосте
  102.         cx_h[i] = (float)i;
  103.         cy_h[i] = 0.87f;
  104.     }
  105.  
  106.     //cublasHandle_t тип указателя на непрозрачную структуру, содержащую контекст библиотеки cuBLAS. Контекст библиотеки cuBLAS должен быть
  107.     //инициализирован с использованием cublasCreate () и возвращенный дескриптор должен быть передан всем последующим вызовам
  108.     //библиотечных функций. Контекст должен быть уничтожен в конце, используяcublasDestroy (),
  109.     cublasHandle_t cublas_handle;//создание объекта  для работы с библиотекой
  110.     cublasCreate(&cublas_handle);//иницилазация работы с библиотекой
  111.  
  112.     const int num_rows = N;
  113.     const int num_cols = 1;
  114.     const int stride = 1;
  115.     //const size_t elem_size = sizeof(float);
  116.  
  117.     //Эта функция копирует плитку строки х столбцы элементы из матрицы A в пространстве памяти хоста в матрицу Вв пространстве памяти GPU
  118.     //cublasSetMatrix(num_rows, num_cols, elem_size, cx_h, num_rows, cx_d, num_rows);
  119.     //cublasSetMatrix(num_rows, num_cols, elem_size, cy_h, num_rows, cy_d, num_rows);
  120.  
  121.     // инициализация вектора на девайсе
  122.     сublasSetVector(num_rows, sizeof(float), cx_h, stride, cx_d, stride);
  123.     cublasSetVector(num_rows, sizeof(float), cy_h, stride, cy_d, stride);
  124.  
  125.     const int stride = 1;
  126.     float alpha = 3.0f;
  127.  
  128.     cudaEventRecord(start, 0);
  129.     //Сложение векторов (реальная одинарная точность).
  130.     //cublas_handle - контекст CUBLAS.
  131.     //N - количество элементов во входных векторах.
  132.     //alpha ( numpy.float32 ) - скаляр .
  133.     //cx_d - указатель на входной вектор с одинарной точностью.
  134.     //stride -  Интервал хранения между элементами x .
  135.     //cy_d - указатель на вектор ввода / вывода одинарной точности.
  136.     //stride - Интервал хранения между элементами y .
  137.     cublasSaxpy(cublas_handle, N, &alpha, cx_d, stride, cy_d, stride);
  138.     cudaEventRecord(stop, 0);
  139.  
  140.     cudaEventSynchronize(stop);// ждем выполнения всех нитей для события завершения
  141.     cudaEventElapsedTime(&elapsedTime, start, stop);
  142.  
  143.     //Эта функция копирует плитку строки х столбцы элементы из матрицы A в пространстве памяти GPU до матрицы Вв пространстве памяти хоста
  144.     //cublasGetMatrix(num_rows, num_cols, elem_size, cx_d, num_rows, cx_h, num_rows);
  145.     //cublasGetMatrix(num_rows, num_cols, elem_size, cy_d, num_rows, cy_h, num_rows);
  146.  
  147.     cublasGetVector(num_rows, sizeof(float), cy_d, stride, cy_h, stride);
  148.     printf("cuBLAS Time:\n \t %f ms\n", elapsedTime);
  149.  
  150.     //Эта функция высвобождает аппаратные ресурсы, используемые библиотекой cuBLAS
  151.     cublasDestroy(cublas_handle);
  152.  
  153.     //Освобождает пространство памяти, на которое указывает hostPtr, которое должно быть возвращено предыдущим вызовом cudaMallocHost () или cudaHostAlloc ()
  154.     cudaFreeHost(cx_h);//высвобождение pinned-памяти
  155.     cudaFreeHost(cy_h);
  156.  
  157.     cudaFree(cx_d);// высвобождение хостовой памяти
  158.     cudaFree(cy_d);
  159.  
  160.     cudaEventDestroy(start);
  161.     cudaEventDestroy(stop);
  162. }
  163.  
  164.  
  165. void ThrustLib() {
  166.     float elapsedTime;
  167.     cudaEvent_t start, stop;
  168.     cudaEventCreate(&start);
  169.     cudaEventCreate(&stop);
  170.  
  171.     //контейнеры, параметризованные типом хранящихся в них значений
  172.     //хранит содержимое вектора в памяти CPU
  173.     thrust::host_vector<float> h1(N);// инициализация векторов на хосте
  174.     thrust::host_vector<float> h2(N);
  175.  
  176.     //заполняет диапазон [first, last)последовательностью чисел.
  177.     //Для каждого итератора iв диапазоне [first, last)эта версия sequenceвыполняет присваивание *i = init + (i - first).
  178.     thrust::sequence(h1.begin(), h1.end());// создаем секвенцию в векторе h1 от начала до конца, с заполнением от 1 до N
  179.  
  180.     //fillприсваивает значение value каждому элементу в диапазоне [first, last)
  181.     thrust::fill(h2.begin(), h2.end(), 0.87);// заполнение вектора неким числом
  182.  
  183.     //контейнеры, параметризованные типом хранящихся в них значений
  184.     //хранит содержимое вектора в памяти GPU
  185.     thrust::device_vector<float> d1 = h1;// копируем векторы хоста в векторы на девайс
  186.     thrust::device_vector<float> d2 = h2;
  187.  
  188.     cudaEventRecord(start, 0);//фиксируем событие начала
  189.  
  190.     //см реализацию выше
  191.     saxpy(3.0, d1, d2);
  192.     cudaEventRecord(stop, 0);
  193.  
  194.     cudaEventSynchronize(stop);// синхронизируем ждем
  195.     cudaEventElapsedTime(&elapsedTime, start, stop);// подсчитываем финальный итог
  196.  
  197.    // обратное копирование
  198.     h2 = d2;
  199.     h1 = d1;
  200.  
  201.     printf("Thrust Time:\n \t %f ms\n\n", elapsedTime);
  202.  
  203.     cudaEventDestroy(start);
  204.     cudaEventDestroy(stop);
  205. }
  206.  
  207.  
  208. int main() {
  209.     CudaKernel();
  210.     cudaCublas();
  211.     ThrustLib();
  212.     return 0;
  213. }
RAW Paste Data