Использование CUDA Runtime API для вычислений. Сравнение CPU и GPU вычислений.
В данной статье я решил провести сравнение выполнения алгоритма написанного на C++ на центральном и графическом процессоре(выполнение вычислений с помощью Nvidia CUDA Runtime API на поддерживаемом GPU Nvidia). CUDA API позволяет выполнение некоторых вычислений на графическом процессоре. Файл c++ использующий cuda, будет иметь расширение .cu.
Схема работы алгоритма приведена ниже.
Задача алгоритма состоит в том, что найти возможные числа X, при возведении которых в степень degree_of, будет получатся исходное число max_number. Сразу отмечу, что все числа которые будут передаваться GPU, будут хранится в массивах. Алгоритм, выполняемый каждым потоком, имеет приблизительно следующий вид:
intdegree_of=2;intdegree_of_max=Number_degree_of_max[0];//Массивхранящийзначениемаксимальнойстепеничислаintx=thread;//номервыполняемогопотокаintmax_number=INPUT[0];//Массивхранящийчисло,котороенеобходимополучитьintNumber=1;intDegree;boolBREAK=false;//Переменнаядлязавершенияwhilewhile(degree_of<=degree_of_max&&!BREAK){Number=1;for(inti=0;i<degree_of;i++){Number*=x;Degree=degree_of;}if(Number==max_number){OUT_NUMBER[thread]=X;//OUT_NUMBERМассивхранящийчислакоторыенеобходимовозвестивстепеньDegreeдляполученияисходногочислаOUT_DEGREE[thread]=Degree;//OUT_DEGREEМассивхранящийстепеньвкоторуюнужновозвестичислоXдляполученияисходногочисла}degree_of++;//Вслучаевыходазапредел:if(degree_of>degree_of_max||Number>max_number){BREAK=true;}}
Код для выполнения на CPU
#include <iostream>#include<vector>#include<string>//необходимо для getline#include<thread>#include<fstream>using namespace std;int Running_thread_counter = 0;void Upload_to_CPU(unsigned long long *Number, unsigned long long *Stepn, bool *Stop,unsigned long long *INPUT, unsigned long long *max, int THREAD);void Upload_to_CPU(unsigned long long *Number, unsigned long long *Stepn, bool *Stop,unsigned long long *INPUT, unsigned long long *max, int THREAD) {int thread = THREAD;Running_thread_counter++;unsigned long long MAX_DEGREE_OF = max[0];int X = thread;unsigned long long Calculated_number = 1;unsigned long long DEGREE_OF = 2;unsigned long long INP = INPUT[0];Stop[thread] = false;bool BREAK = false;if (X != 0 && X != 1) {while (!BREAK) {if (DEGREE_OF <= MAX_DEGREE_OF) {Calculated_number = 1;for (int counter = 0; counter < DEGREE_OF; counter++) {Calculated_number *= X;}if (Calculated_number == INP) {Stepn[thread] = DEGREE_OF;Number[thread] = X;Stop[thread] = true;BREAK = true;}DEGREE_OF++;}else { BREAK = true; }}}}void Parallelize_to_threads(unsigned long long *Number, unsigned long long *Stepn, bool *Stop,unsigned long long *INPUT, unsigned long long *max, int size);int main(){int size = 1000;unsigned long long *Number = new unsigned long long[size], *Degree_of = new unsigned long long[size];unsigned long long *Max_Degree_of = new unsigned long long[1];unsigned long long *INPUT_NUMBER = new unsigned long long[1];Max_Degree_of[0] = 7900;INPUT_NUMBER[0] = 216 * 216 * 216;ifstream inp("input.txt");if (inp.is_open()) {string t;vector<unsigned long long>IN;while (getline(inp, t)) {IN.push_back(stol(t));}INPUT_NUMBER[0] = IN[0];//исходное числоMax_Degree_of[0] = IN[1];//значение максимальной степени}else {ofstream error("error.txt");if (error.is_open()) {error << "No file " << '"' << "input.txt" << '"' << endl;error << "Please , create a file" << '"' << "input.txt" << '"' << endl;error << "One read:input number" << endl;error << "Two read:input max stepen" << endl;error << "." << endl;error.close();INPUT_NUMBER[0] = 1;Max_Degree_of[0] = 1;}}//расскометрируйте следующий код , если хотите видеть исходные значения в окне консоли //cout << INPUT[0] << endl;bool *Elements_that_need_to_stop = new bool[size];Parallelize_to_threads(Number, Degree_of, Elements_that_need_to_stop, INPUT_NUMBER, Max_Degree_of, size);vector<unsigned long long>NUMBER, DEGREEOF;for (int i = 0; i < size; i++) {if (Elements_that_need_to_stop[i]) {if (Degree_of[i] < INPUT_NUMBER[0] && Number[i] < INPUT_NUMBER[0]) {//проверка на ошибки NUMBER.push_back(Number[i]);DEGREEOF.push_back(Degree_of[i]);}}}//расскометрируйте следующий код , если хотите вывести результаты в консоль//это может замедлить программу /*for (int f = 0; f < NUMBER.size(); f++) {cout << NUMBER[f] << "^" << DEGREEOF[f] << "=" << INPUT_NUMBER[0] << endl;}*/ofstream out("out.txt");if (out.is_open()) {for (int f = 0; f < NUMBER.size(); f++) {out << NUMBER[f] << "^" << DEGREEOF[f] << "=" << INPUT_NUMBER[0] << endl;}out.close();}}void Parallelize_to_threads(unsigned long long *Number, unsigned long long *Stepn, bool *Stop,unsigned long long *INPUT, unsigned long long *max, int size) {thread *T = new thread[size];Running_thread_counter = 0;for (int i = 0; i < size; i++) {T[i] = thread(Upload_to_CPU, Number, Stepn, Stop, INPUT, max, i);T[i].detach();}while (Running_thread_counter < size - 1);//дождаться завершения выполнения всех потоков }
Для работы алгоритма необходим текстовый файл с исходным числом и максимальной степенью.
Код для выполнения вычислений на GPU
//библиотеки cuda_runtime.h и device_launch_parameters.h//для работы с cyda#include "cuda_runtime.h"#include "device_launch_parameters.h"#include<vector>#include<string>//для getline#include <stdio.h>#include<fstream>using namespace std;__global__ void Upload_to_GPU(unsigned long long *Number,unsigned long long *Stepn, bool *Stop,unsigned long long *INPUT,unsigned long long *max) {int thread = threadIdx.x;unsigned long long MAX_DEGREE_OF = max[0]; int X = thread;unsigned long long Calculated_number = 1;unsigned long long Current_degree_of_number = 2; unsigned long long Original_numberP = INPUT[0];Stop[thread] = false;bool BREAK = false;if (X!=0&&X!=1) {while (!BREAK) {if (Current_degree_of_number <= MAX_DEGREE_OF) {Calculated_number = 1;for (int counter = 0; counter < Current_degree_of_number; counter++) { Calculated_number*=X;}if (Calculated_number == Original_numberP) {Stepn[thread] = Current_degree_of_number;Number[thread] = X;Stop[thread] = true;BREAK = true;}Current_degree_of_number++;}else { BREAK = true; }}}}cudaError_t Configure_cuda(unsigned long long *Number, unsigned long long *Stepn, bool *Stop,unsigned long long *INPUT, unsigned long long *max,unsigned int size);int main(){int size = 1000; unsigned long long *Number=new unsigned long long [size], *Degree_of=new unsigned long long [size];unsigned long long *Max_degree_of = new unsigned long long [1];unsigned long long *INPUT_NUMBER = new unsigned long long [1]; Max_degree_of[0] = 7900;ifstream inp("input.txt");if (inp.is_open()) {string text;vector<unsigned long long>IN;while (getline(inp, text)) {IN.push_back( stol(text));}INPUT_NUMBER[0] = IN[0];Max_degree_of[0] = IN[1];}else {ofstream error("error.txt");if (error.is_open()) {error<<"No file "<<'"'<<"input.txt"<<'"'<<endl;error<<"Please , create a file" << '"' << "input.txt" << '"' << endl;error << "One read:input number" << endl;error << "Two read:input max stepen" << endl;error << "." << endl;error.close();INPUT_NUMBER[0] = 1;Max_degree_of[0] = 1;}}bool *Elements_that_need_to_stop = new bool[size]; // Загрузка массивов в cudacudaError_t cudaStatus = Configure_cuda(Number, Degree_of, Elements_that_need_to_stop, INPUT_NUMBER, Max_degree_of, size); if (cudaStatus != cudaSuccess) { fprintf(stderr, "addWithCuda failed!"); return 1; }vector<unsigned long long>NUMBER, DEGREEOF;for (int i = 0; i < size; i++) {if (Elements_that_need_to_stop[i]) {NUMBER.push_back(Number[i]);//занести в вектор числоDEGREEOF.push_back(Degree_of[i]);//занести в вектор степень числа}}//раскоментируйте следующий код , чтобы вывести результаты в консоль/*for (int f = 0; f < NUMBER.size(); f++) {cout << NUMBER[f] << "^" << DEGREEOF[f] << "=" << INPUT_NUMBER[0] << endl;}*/ofstream out("out.txt");if (out.is_open()) {for (int f = 0; f < NUMBER.size(); f++) {out << NUMBER[f] << "^" << DEGREEOF[f] << "=" << INPUT_NUMBER[0] << endl;}out.close();} //Очистить ресурсы связанные с устройством cudaStatus = cudaDeviceReset(); if (cudaStatus != cudaSuccess) { fprintf(stderr, "cudaDeviceReset failed!"); return 1; } return 0;}cudaError_t Configure_cuda(unsigned long long *Number, unsigned long long *Degree_of, bool *Stop,unsigned long long *INPUT, unsigned long long *max,unsigned int size) {unsigned long long *dev_Number = 0;unsigned long long *dev_Degree_of = 0;unsigned long long *dev_INPUT = 0;unsigned long long *dev_Max = 0;bool *dev_Elements_that_need_to_stop;cudaError_t cudaStatus;// УСТАНОВКА ИСПОЛЬЗУЕМОГО GPU cudaStatus = cudaSetDevice(0);if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaSetDevice failed! Do you have a CUDA-capable GPU installed?");goto Error;}// РЕЗЕРВИРОВАНИЕ МЕСТА В ПАМЯТИ ПОД ДАННЕcudaStatus = cudaMalloc((void**)&dev_Number, size * sizeof(unsigned long long));if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMalloc failed!dev_Number");goto Error;}cudaStatus = cudaMalloc((void**)&dev_Degree_of, size * sizeof(unsigned long long));if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMalloc failed!dev_Degree_of");goto Error;}cudaStatus = cudaMalloc((void**)&dev_Max, size * sizeof(unsigned long long int));if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMalloc failed!dev_Max");goto Error;}cudaStatus = cudaMalloc((void**)&dev_INPUT, size * sizeof(unsigned long long));if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMalloc failed!dev_INPUT");goto Error;}cudaStatus = cudaMalloc((void**)&dev_Elements_that_need_to_stop, size * sizeof(bool));if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMalloc failed!dev_Stop");goto Error;}// ПЕРЕМЕЩЕНИЕ ДАННХ В ПАМЯТЬ GPUcudaStatus = cudaMemcpy(dev_Max, max, size * sizeof(unsigned long long), cudaMemcpyHostToDevice);if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMemcpy failed!");goto Error;}cudaStatus = cudaMemcpy(dev_INPUT, INPUT, size * sizeof(unsigned long long), cudaMemcpyHostToDevice);if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMemcpy failed!");goto Error;}Upload_to_GPU<<<1, size>>>(dev_Number, dev_Degree_of, dev_Elements_that_need_to_stop, dev_INPUT, dev_Max);// Проверка сбоев ядраcudaStatus = cudaGetLastError();if (cudaStatus != cudaSuccess) {fprintf(stderr, "addKernel launch failed: %s\n", cudaGetErrorString(cudaStatus));goto Error;}// Ожидание завершения операций , выполняемых ядромcudaStatus = cudaDeviceSynchronize();if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaDeviceSynchronize returned error code %d after launching addKernel!\n", cudaStatus);goto Error;}// Перемещение данных из памяти GPU в системную памятьcudaStatus = cudaMemcpy(Number, dev_Number, size * sizeof(unsigned long long), cudaMemcpyDeviceToHost);if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMemcpy failed!");goto Error;}cudaStatus = cudaMemcpy(Degree_of, dev_Degree_of, size * sizeof(unsigned long long), cudaMemcpyDeviceToHost);if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMemcpy failed!");goto Error;}cudaStatus = cudaMemcpy(Stop, dev_Elements_that_need_to_stop, size * sizeof(bool), cudaMemcpyDeviceToHost);if (cudaStatus != cudaSuccess) {fprintf(stderr, "cudaMemcpy failed!");goto Error;}Error://Освобождение памяти GPU от данныхcudaFree(dev_INPUT);cudaFree(dev_Degree_of);cudaFree(dev_Max);cudaFree(dev_Elements_that_need_to_stop);cudaFree(dev_Number);return cudaStatus;}
Идентификатор
__global__
в .cu файле указывает, что функция выполняется на уровне
ядра GPU.Для работы с cyda, перед вызовом функции, нужно зарезервировать память под массив и перенести элементы в память GPU. Это увеличивает объем кода, но позволяет разгрузить CPU, так как вычисления производятся на GPU.Поэтому ,cuda, дает как минимум возможность разгрузить процессор для выполнения других нагрузок, не использующих cuda.
В случае примера на cuda, задача процессора заключается лишь в загрузке инструкций на GPU и обработке результатов пришедших с GPU; В то время как в коде для CPU, процессор обрабатывает каждый поток. Стоит отметить, что cyda имеет ограничения по количеству запускаемых потоков, поэтому в обоих алгоритмах я взял одинаковое количество потоков, равное 1000. Также, в случае с CPU я использовал переменную
intRunning_thread_counter=0;
чтобы считать количество уже выполненных потоков и дожидаться, пока все потоки не выполнятся.
Тестируемая конфигурация
Сведения о cyda были взяты из GPU-Z
- CPU :amd ryzen 5 1400(4core,8thread)
- ОЗУ:8гбDDR4 2666
- GPU:Nvidia rtx 2060
- OS:windows 10 version 2004
- Cuda:
- Compute Capability 7.5
- Threads per Multiprocessor 1024
- CUDA 11.1.70
- GPU-Z:version 2.35.0
- Visual Studio 2017
Сведения о cyda были взяты из GPU-Z
Для тестирования алгоритма я использовал
следующий код на C#
, который создавал файл с исходными данными, затем последовательно
запускал exe файлы алгоритмов использующих CPU или GPU и замерял
время их работы, затем заносил это время и результаты работы
алгоритмов в файл result.txt. Для замера загруженности
процессора использовался диспетчер задач windows.
usingSystem;usingSystem.Collections.Generic;usingSystem.Linq;usingSystem.Text;usingSystem.Threading.Tasks;usingSystem.Diagnostics;usingSystem.IO;namespaceConsoleAppTESTSTEPEN_CPU_AND_GPU_{classProgram{staticstringUpload(Int64number,Int64degree_of){stringOUT="";string[]Chord_values=newstring[2];Int64Degree_of=degree_of;Int64Number=number;Chord_values[0]=Number.ToString();Chord_values[1]=Degree_of.ToString();File.WriteAllLines("input.txt",Chord_values);//файлвходныхданныхOUT+="inputnumber:"+Number.ToString()+"\n";OUT+="inputdegreeofnumber:"+Degree_of.ToString()+"\n";DateTimerunning_CPU_application=DateTime.Now;//записатьвремязапускапрограммыProcessproc=Process.Start("ConsoleApplication29.exe");//exeреализацияалгоритманаc++x64использующаяCPUдлявычисленийwhile(!proc.HasExited);//дождатсязавершенияпрограммыDateTimestop_CPU_application=DateTime.Now;//записатьвремяостановкипрограммыstring[]outs=File.ReadAllLines("out.txt");//получитьрезультатыFile.Delete("out.txt");OUT+="CPU:"+"\n";if(outs.Length>0){for(intj=0;j<outs.Length;j++){OUT+=outs[j]+"\n";}}else{OUT+="novalues"+"\n";}OUT+="running_CPU_application:"+running_CPU_application.ToString()+"\n";OUT+="stop_CPU_application:"+stop_CPU_application.ToString()+"\n";OUT+="GPU:"+"\n";//альтернативныедействиядляреализацииалгоритмаkorenXN.exex64использующегодлявычисленийGPUDateTimerunning_GPU_application=DateTime.Now;ProcessprocGPU=Process.Start("korenXN.exe");while(!procGPU.HasExited);DateTimestop_GPU_application=DateTime.Now;string[]outs2=File.ReadAllLines("out.txt");File.Delete("out.txt");if(outs2.Length>0){for(intj=0;j<outs2.Length;j++){OUT+=outs2[j]+"\n";}}else{OUT+="novalues"+"\n";}OUT+="running_GPU_application:"+running_GPU_application.ToString()+"\n";OUT+="stop_GPU_application:"+stop_GPU_application.ToString()+"\n";returnOUT;//возвратитьрезультат}staticvoidMain(){Int64start=36*36;//начальноезначениевходногочислаInt64degree_of_strat=500;//начальноезначениемаксимальнойстепениintsize=20-5;//количествоэлементоввмассивеInt64[]Number=newInt64[size];//массиввходныхчиселInt64[]Degree_of=newInt64[size];//массивмаксимальныхстепенейstring[]outs=newstring[size];//масссиврезультатовfor(intn=0;n<size;n++){if(n%2==0){Number[n]=start*start;}else{Number[n]=start*degree_of_strat;Number[n]-=n+n;}start+=36*36;Degree_of[n]=degree_of_strat;degree_of_strat+=1000;}for(intn=0;n<size;n++){outs[n]=Upload(Number[n],Degree_of[n]);Console.WriteLine(outs[n]);}System.IO.File.WriteAllLines("result.txt",outs);//записатьрезультатывфайлresult.txt}}}
Результаты теста превидены в таблице:
Как видно из таблицы, время выполнения алгоритма на GPU немного больше, чем на CPU.
Однако, отмечу, что вовремя работы алгоритма использующего для вычислений GPU загрузка им CPU, в Диспетчере задач, не превышала 30%, в то время как алгоритм использующий для вычислений CPU, загружал его на 68-85%, что в свою очередь иногда приводило к замедлению других приложений. Также, ниже приведен график, показывающий различие во
времени выполнения (по оси Y)CPU и GPU в зависимости от входного числа(по оси X).
график
Далее я решил провести тестирование при процессоре нагруженном другими приложениями. Процессор был нагружен так, что запущенный в приложение тест, не занимал больше 55% ресурсов процессора. Результаты теста приведены ниже:
График
Как видно из таблицы, в случае с нагруженным CPU, выполнение вычислений на GPU, дает прирост производительности, так как загруженность процессора в 30% укладывается в лимит 55%, а в случае использования CPU для вычислений, его загрузка составляет 68-85% , что тормозит работу алгоритма, если CPU нагружен другими приложениями.
Поэтому, можно сделать вывод, что использование GPU для вычислений, не обязательно должно давать более быструю работу алгоритма, однако, оно способно разгрузить CPU, что может играть роль, если он нагружен другими приложениями.
