Skip to content

Commit

Permalink
Update main.cpp
Browse files Browse the repository at this point in the history
  • Loading branch information
@Smart781
Smart781 authored Dec 11, 2024
1 parent e32dfbe commit 5d45bb5
Showing 1 changed file with 72 additions and 17 deletions.
89 changes: 72 additions & 17 deletions src/main.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -40,17 +40,51 @@ int main() {
// TODO 1 По аналогии с предыдущим заданием узнайте, какие есть устройства, и выберите из них какое-нибудь
// (если в списке устройств есть хоть одна видеокарта - выберите ее, если нету - выбирайте процессор)

cl_uint platformCount;
OCL_SAFE_CALL(clGetPlatformIDs(0, nullptr, &platformCount));
std::vector<cl_platform_id> platforms(platformCount);
OCL_SAFE_CALL(clGetPlatformIDs(platformCount, platforms.data(), nullptr));

cl_device_id chosenDevice = nullptr;
for (const auto& platform : platforms) {
cl_uint deviceCount;
OCL_SAFE_CALL(clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 0, nullptr, &deviceCount));
std::vector<cl_device_id> devices(deviceCount);
OCL_SAFE_CALL(clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, deviceCount, devices.data(), nullptr));

for (const auto& device : devices) {
cl_device_type deviceType;
OCL_SAFE_CALL(clGetDeviceInfo(device, CL_DEVICE_TYPE, sizeof(deviceType), &deviceType, nullptr));

if (deviceType & CL_DEVICE_TYPE_GPU) {
chosenDevice = device;
break;
} else if (chosenDevice == nullptr) {
chosenDevice = device;
}
}
if (chosenDevice) break;
}
assert(chosenDevice != nullptr);

// TODO 2 Создайте контекст с выбранным устройством
// См. документацию https://www.khronos.org/registry/OpenCL/sdk/1.2/docs/man/xhtml/ -> OpenCL Runtime -> Contexts -> clCreateContext
// Не забывайте проверять все возвращаемые коды на успешность (обратите внимание, что в данном случае метод возвращает
// код по переданному аргументом errcode_ret указателю)

// Контекст и все остальные ресурсы следует освобождать с помощью clReleaseContext/clReleaseQueue/clReleaseMemObject... (да, не очень RAII, но это лишь пример)

cl_int err;
cl_context context = clCreateContext(nullptr, 1, &chosenDevice, nullptr, nullptr, &err);
OCL_SAFE_CALL(err);

// TODO 3 Создайте очередь выполняемых команд в рамках выбранного контекста и устройства
// См. документацию https://www.khronos.org/registry/OpenCL/sdk/1.2/docs/man/xhtml/ -> OpenCL Runtime -> Runtime APIs -> Command Queues -> clCreateCommandQueue
// Убедитесь, что в соответствии с документацией вы создали in-order очередь задач

cl_command_queue queue = clCreateCommandQueue(context, chosenDevice, 0, &err);
OCL_SAFE_CALL(err);

unsigned int n = 1000 * 1000;
// Создаем два массива псевдослучайных данных для сложения и массив для будущего хранения результата
std::vector<float> as(n, 0);
Expand All @@ -69,6 +103,13 @@ int main() {
// Данные в as и bs можно прогрузить этим же методом, скопировав данные из host_ptr=as.data() (и не забыв про битовый флаг, на это указывающий)
// или же через метод Buffer Objects -> clEnqueueWriteBuffer

cl_mem as_gpu = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, n * sizeof(float), as.data(), &err);
OCL_SAFE_CALL(err);
cl_mem bs_gpu = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, n * sizeof(float), bs.data(), &err);
OCL_SAFE_CALL(err);
cl_mem cs_gpu = clCreateBuffer(context, CL_MEM_WRITE_ONLY, n * sizeof(float), nullptr, &err);
OCL_SAFE_CALL(err);

// TODO 6 Выполните TODO 5 (реализуйте кернел в src/cl/aplusb.cl)
// затем убедитесь, что выходит загрузить его с диска (убедитесь что Working directory выставлена правильно - см. описание задания),
// напечатав исходники в консоль (if проверяет, что удалось считать хоть что-то)
Expand All @@ -86,9 +127,15 @@ int main() {
// см. Runtime APIs -> Program Objects -> clCreateProgramWithSource
// у string есть метод c_str(), но обратите внимание, что передать вам нужно указатель на указатель

const char* kernel_sources_cstr = kernel_sources.c_str();
cl_program program = clCreateProgramWithSource(context, 1, &kernel_sources_cstr, nullptr, &err);
OCL_SAFE_CALL(err);

// TODO 8 Теперь скомпилируйте программу и напечатайте в консоль лог компиляции
// см. clBuildProgram

OCL_SAFE_CALL(clBuildProgram(program, 1, &chosenDevice, nullptr, nullptr, nullptr));

// А также напечатайте лог компиляции (он будет очень полезен, если в кернеле есть синтаксические ошибки - т.е. когда clBuildProgram вернет CL_BUILD_PROGRAM_FAILURE)
// Обратите внимание, что при компиляции на процессоре через Intel OpenCL драйвер - в логе указывается, какой ширины векторизацию получилось выполнить для кернела
// см. clGetProgramBuildInfo
Expand All @@ -102,13 +149,16 @@ int main() {
// TODO 9 Создайте OpenCL-kernel в созданной подпрограмме (в одной подпрограмме может быть несколько кернелов, но в данном случае кернел один)
// см. подходящую функцию в Runtime APIs -> Program Objects -> Kernel Objects

cl_kernel kernel = clCreateKernel(program, "aplusb", &err);
OCL_SAFE_CALL(err);

// TODO 10 Выставите все аргументы в кернеле через clSetKernelArg (as_gpu, bs_gpu, cs_gpu и число значений, убедитесь, что тип количества элементов такой же в кернеле)
{
// unsigned int i = 0;
// clSetKernelArg(kernel, i++, ..., ...);
// clSetKernelArg(kernel, i++, ..., ...);
// clSetKernelArg(kernel, i++, ..., ...);
// clSetKernelArg(kernel, i++, ..., ...);
unsigned int i = 0;
OCL_SAFE_CALL(clSetKernelArg(kernel, i++, sizeof(cl_mem), &as_gpu));
OCL_SAFE_CALL(clSetKernelArg(kernel, i++, sizeof(cl_mem), &bs_gpu));
OCL_SAFE_CALL(clSetKernelArg(kernel, i++, sizeof(cl_mem), &cs_gpu));
OCL_SAFE_CALL(clSetKernelArg(kernel, i++, sizeof(unsigned int), &n));
}

// TODO 11 Выше увеличьте n с 1000*1000 до 100*1000*1000 (чтобы дальнейшие замеры были ближе к реальности)
Expand All @@ -125,8 +175,10 @@ int main() {
size_t global_work_size = (n + workGroupSize - 1) / workGroupSize * workGroupSize;
timer t;// Это вспомогательный секундомер, он замеряет время своего создания и позволяет усреднять время нескольких замеров
for (unsigned int i = 0; i < 20; ++i) {
// clEnqueueNDRangeKernel...
// clWaitForEvents...
cl_event event;
OCL_SAFE_CALL(clEnqueueNDRangeKernel(queue, kernel, 1, nullptr, &global_work_size, &workGroupSize, 0, nullptr, &event));
OCL_SAFE_CALL(clWaitForEvents(1, &event));
clReleaseEvent(event);
t.nextLap();// При вызове nextLap секундомер запоминает текущий замер (текущий круг) и начинает замерять время следующего круга
}
// Среднее время круга (вычисления кернела) на самом деле считается не по всем замерам, а лишь с 20%-перцентайля по 80%-перцентайль (как и стандартное отклонение)
Expand All @@ -140,34 +192,37 @@ int main() {
// - Флопс - это число операций с плавающей точкой в секунду
// - В гигафлопсе 10^9 флопсов
// - Среднее время выполнения кернела равно t.lapAvg() секунд
std::cout << "GFlops: " << 0 << std::endl;
double gflops = (n / t.lapAvg()) / 1e9;
std::cout << "GFlops: " << gflops << std::endl;

// TODO 14 Рассчитайте используемую пропускную способность обращений к видеопамяти (в гигабайтах в секунду)
// - Всего элементов в массивах по n штук
// - Размер каждого элемента sizeof(float)=4 байта
// - Обращений к видеопамяти 2*n*sizeof(float) байт на чтение и 1*n*sizeof(float) байт на запись, т.е. итого 3*n*sizeof(float) байт
// - В гигабайте 1024*1024*1024 байт
// - Среднее время выполнения кернела равно t.lapAvg() секунд
std::cout << "VRAM bandwidth: " << 0 << " GB/s" << std::endl;
double bandwidth = (3.0 * n * sizeof(float) / (1024 * 1024 * 1024)) / t.lapAvg();
std::cout << "VRAM bandwidth: " << bandwidth << " GB/s" << std::endl;
}

// TODO 15 Скачайте результаты вычислений из видеопамяти (VRAM) в оперативную память (RAM) - из cs_gpu в cs (и рассчитайте скорость трансфера данных в гигабайтах в секунду)
{
timer t;
for (unsigned int i = 0; i < 20; ++i) {
// clEnqueueReadBuffer...
OCL_SAFE_CALL(clEnqueueReadBuffer(queue, cs_gpu, CL_TRUE, 0, n * sizeof(float), cs.data(), 0, nullptr, nullptr));
t.nextLap();
}
std::cout << "Result data transfer time: " << t.lapAvg() << "+-" << t.lapStd() << " s" << std::endl;
std::cout << "VRAM -> RAM bandwidth: " << 0 << " GB/s" << std::endl;
double transfer_bandwidth = (n * sizeof(float) / (1024 * 1024 * 1024)) / t.lapAvg();
std::cout << "VRAM -> RAM bandwidth: " << transfer_bandwidth << " GB/s" << std::endl;
}

// TODO 16 Сверьте результаты вычислений со сложением чисел на процессоре (и убедитесь, что если в кернеле сделать намеренную ошибку, то эта проверка поймает ошибку)
// for (unsigned int i = 0; i < n; ++i) {
// if (cs[i] != as[i] + bs[i]) {
// throw std::runtime_error("CPU and GPU results differ!");
// }
// }

for (unsigned int i = 0; i < n; ++i) {
if (cs[i] != as[i] + bs[i]) {
throw std::runtime_error("CPU and GPU results differ!");
}
}
return 0;
}

0 comments on commit 5d45bb5

Please sign in to comment.