Что за программа opencl

Содержание

Одна из главных проблем взаимодействия технологий и пользователей — проблема совместимости. Зачастую некоторые элементы системы отказываются работают совместно друг с другом, из-за чего возникают трудности в интеграции, требующие сложных решений.

Вопреки лозунгу технологии «plug-and-play» («подключил-и-работай»), принтеры могут не подружиться с ПК, серверы — не всегда корректно взаимодействуют с серверами, работающими под управлением другой операционной системы, приложения разных разработчиков плохо уживаются в одной системе, мобильные устройства не синхронизируются с «большим братьями»… Список можно продолжать почти до бесконечности. Главная сложность даже не во взаимодействии аппаратного и программного обеспечения как таковом, а в том, что постоянно происходят значительные изменения в технологиях. Особенно — в архитектуре центральных и графических процессоров. А разработчикам и пользователям трудно принять эти перемены.

Так, например, в последнее время значительно возрос интерес к использованию графических процессоров для проведения вычислений общего назначения. Конвейеры графической обработки эволюционировали из модулей, выполняющих фиксированный набор функций, до устройств, использующих полностью программируемые вычислительные блоки. Видя это, находчивые разработчики пытались использовать грандиозный потенциал параллельных вычислений на графических ядрах, переписывая свои приложения так, чтобы вписываться в модели программирования, поддерживаемые традиционными программными интерфейсами графических приложений, такими как Microsoft DirectX и OpenGL.

Episode 1: What is OpenCL™?

Хотя такой подход снискал популярность среди исследователей, но оказался труден для разработчиков, не привыкших иметь дело с графическими API. Это привело к разработке новых программных интерфейсов, предоставляющих разработчикам больше удобств за счёт скрытия деталей модели графического программирования. Однако массовый переход на новые интерфейсы всё ещё был затруднён.

Дело в том, что создаваемые API оказывались собственническими, работая на продуктах только поставщика интерфейса. Кроме того, графический процессор рассматривался как независимое устройство, позволяющее ускорить вычисления, а не как составная часть сбалансированной однородной платформы. В результате, для GPU и CPU писали совершенно разный код — каждому отдавалась определённая часть функций. В свою очередь, это приводило к медленному принятию нового подхода из-за растерянности разработчиков и маркетологов, пытавшихся найти верный путь полного использования дополнительной вычислительной мощи и понять, какой программный интерфейс лучше использовать для модуля.

История показывает, что рыночная путаница стандартов способна сильно повлиять на продажи продуктов – достаточно вспомнить противостояние стандартов VHS и Betamax или Blu-ray и HD DVD. Пользователи, будь то корпоративные или домашние, могут очень неохотно инвестировать в технологию, если её перспективы туманны.

Производители вскоре поняли, что и принятие технологии вычисления на графических процессорах тормозило именно отсутствие единого стандарта. И это осознание подтолкнуло их к утверждению стандарта и началу активного сотрудничества. В прошлом году был образован некоммерческий консорциум Khronos Group, объединивший таких крупных поставщиков CPU и GPU, как Apple, AMD, Intel, nVidia, Sun Microsystems и Sony. Эти компании объединили свои усилия для создания универсального языка программирования Open Computing Language (OpenCL). Разработчики могут использовать этот новый стандартизированный язык для программирования на базе однородных платформ, используя все доступные в системе процессоры, независимо от их типа и их поставщиков.

Созданный Khronos Group стандарт OpenCL рассчитан на выполнение параллельных вычислений на графических и центральных процессорах. Он позволяет передать часть расчётов от CPU к GPU, сбалансировав нагрузку и — как следствие — увеличив эффективность платформы в целом.

9 декабря 2008 года Khronos Group утвердила стандарт OpenCL 1.0, разработанный компанией Apple. Эту спецификацию поддержали AMD, nVidia, IBM, Intel и другие компании. AMD встроила совместимый с OpenCL компилятор в свой бесплатный пакет ATI Stream SDK для разработчиков; NVIDIA также добавила поддержку этого стандарта параллельных вычислений в свой пакет GPU Computing Toolkit.
20 апреля 2009 года nVidia представила бета-версию драйвера и SDK с поддержкой OpenCL.
5 апреля 2009 года AMD открыла для свободной загрузки beta-версию комплекта разработчика ATI Stream SDK v2.0 в который входит язык мультипроцессорного программирования OpenCL.

OpenCL позволяет разработчикам переносить на GPU выполнение задач, параллельных по своей природе, возлагая выполнение остальных задач на CPU. Каждый процессор в системе работает над теми задачами, для которых он лучше всего подходит, что способствует получению больших результатов за меньшее время на том же аппаратном обеспечении.
C помощью SDK от AMD разработчики уже сейчас могут начать переносить свои программы на OpenCL и пользоваться преимуществами вычислений на графических процессорах, когда выйдет полная платформа разработки OpenCL.

На практике польза от применения OpenCL очевидна. К примеру, он даёт возможность улучшить время отклика серверов и высокопроизводительных систем. Скажем, многие компании, имеющие дело с огромными объёмами данных, часто вынужден тратить огромное время на работу с массивами информации, поскольку мощности CPU их серверов не безграничны, а поток данных иногда нарастает лавинообразно. За счёт же использования вычислительной мощи графических процессоров можно получать желаемые результаты значительно быстрее. Используя все доступные *PU, OpenCL позволяет перераспределить нагрузку, выжимая из уже имеющегося оборудования больше вычислительной мощи.
Впрочем, преимуществами OpenCL могут пользоваться не только крупные компании, а любые организации, проводящие массовую обработку данных. Например, научно-исследовательские лаборатории, университеты, отделы автоматизированного проектирования, ТВ-студии и кинокомпании.

Наглядно убедиться в эффективности использования технологии OpenCL при отрисовке видеоэффектов на 24-х ядерной системе можно, посмотрев видеоролик.

Язык OpenCL выделяется тем, что позволяет приложениям динамически определять, какие *PU и ядра доступны, в результате чего разработчики могут динамически масштабировать производительность своих приложений в зависимости от доступного аппаратного обеспечения клиентов. Например, если компания приобретает приложение с поддержкой ускорения OpenCL, которое может использовать доступные в данный момент центральные и графические процессоры, то она может подогнать инвестиции в оборудование под текущие вычислительные потребности. Если затем её потребности возрастут, то она может просто купить дополнительные или более мощные центральные или графические процессоры.
Возможно, самое важное преимущество стандарта OpenCL состоит в том, что он является открытым. Этот стандарт разрабатывается группой компаний, и все поставщики, вовлечённые в эту разработку, заинтересованы в том, чтобы он работал на их оборудовании. Единый проект позволяет всем его участникам договориться друг с другом и создать язык, подходящий для любого оборудования, от чего также выиграют и конечные пользователи, которые смогут быстрее и эффективнее проводить свои вычисления.
Стандарт OpenCL нацелен на то, чтобы решать реальные проблемы конечных пользователей и программистов. Открытый межплатформенный язык, такой как OpenCL, поможет нам приблизиться к эре быстрых и эффективных вычислений.

Вычисляем на видеокартах. Технология OpenCL. Часть 1b. Пишем для OpenCL

22 июня автор курса «Разработчик C++» в Яндекс.Практикуме Георгий Осипов провёл вебинар «Вычисляем на видеокартах. Технология OpenCL».

0. Зачем мы здесь собрались. Краткая история GPGPU
1a. Как работает OpenCL
1b. Пишем для OpenCL
2. Алгоритмы в условиях массового параллелизма
3. Сравнение технологий

У программы для OpenCL есть две части: kernel-код и host-код — то, что выполняется на видеокарте, и то, что выполняется на компьютере. Кроме того, программу нужно скомпилировать и запустить. Всё это будет рассмотрено в сегодняшней статье. Начнём с самого интересного — напишем часть kernel.

В предыдущих сериях

Kernel-код — код, который выполняется на устройстве, поддерживающем OpenCL: видеокарте, процессоре или другом устройстве.
Host-код — код, который выполняется на центральном процессоре и отдаёт команду на запуск kernel-кода.
Work-item — один поток выполнения, отвечающий за одну элементарную подзадачу большой задачи. Все work-item’ы одной задачи исполняют один и тот же kernel-код.
Warp, или wavefront — группа work-item’ов, выполняющая команды синхронно.
Work group, или рабочая группа, — группа work-item’ов, имеющая общую локальную память и средства синхронизации между потоками.
Локальный размер — размер рабочей группы, измеряемый в work-item’ах. А именно длина, высота (для двумерных и трёхмерных задач) и глубина (для трёхмерных задач).
Глобальный размер — размер всей задачи, измеряемый в work-item’ах.

Kernel

рекурсию,
указатели на функции,
массивы с переменным размером,
стандартные заголовки (stdlib.h, …),
extern , static , auto , register .

Хорошая новость: у OpenCL есть не только ограничения, но и дополнения. Например, очень крутые векторные типы. И честно вам скажу: когда я начал программировать на OpenCL, их стало очень не хватать в обычном программировании на CPU.

С векторными типами можно сделать переменную, в которой будет храниться сразу 16 значений float . 16 — это довольно много, чаще используют два или три значения — так представляют координаты точки. Конечно, вы скажете, что в C++ можно сделать себе какой угодно векторный тип и определить для него все операции. Но я отвечу: не все. Тернарную операцию перегрузить не получится. А в OpenCL она прекрасно работает:

int a = . b = . c = . ; // Инициализируем векторную переменную: int3 coordinates = (int3)(a, b, c); if (all(coordinates > 0)) < // Все координаты положительны. >else if (any(coordinates > 0)) < // Хотя бы одна координата положительна. >// Вычитаем из каждой координаты 100 и сравниваем с нулём. // sign — это вектор, имеющий элементы 1 и -1. // Тут применяется векторная тернарная операция. int3 sign = (coordinates — 100) >= 0 ? 1 : -1; // Можно выбрать из трёх координат любой набор // и сохранить в отдельную переменную: int2 drop_y = coordinates.xz;

Но не будем отвлекаться и вернёмся к написанию kernel. Мы будем писать функцию, рисующую множество Мандельброта.

Пишем точку входа — kernel-функцию. Она всегда void и будет вызвана для каждого work-item’а

Зелёным выделено название kernel, синим — встроенные функции OpenCL, красным — функции, которые мы скоро напишем.

Это двумерная задача. И я хочу, чтобы один work-item рисовал какую-то точку множества Мандельброта. Каждый work-item выполняет код функции независимо от других, и ему прежде всего нужно понять, какую точку множества Мандельброта отрисовать. Чтобы это как-то кастомизировать, в аргументах задаём точку плоскости, которая находится в центре нашей картинки, и масштаб. Разберём аргументы подробнее:

float px , float py — координаты центра отрисовываемой области;
float mag — коэффициент увеличения;
float max_iters — точность прорисовки множества, которая показывает, сколько шагов нужно сделать для каждой точки;
int w , int h — размеры рисуемого изображения;
__global uint* result — память, куда нужно записать ответ;
int result_step — смещение между строками ответа.

Параметр result имеет тип __global uint* , то есть это указатель на глобальную память. При написании host-кода её нужно заранее выделить. После работы kernel, который заполнит эту память, мы скопируем данные в RAM и обработаем их уже на стороне хоста. Тип uint , имеющий размер 4 байта, выбран для записи цвета одного пикселя. В нём 3 байта задают компоненты R, G и B, а один байт не используется.

Последний параметр result_step нужен для корректного заполнения памяти result . Если вы когда-либо программировали обработку изображений, то наверняка знаете, что это практически обязательный атрибут при передаче картинок. result_step ещё называют отступом. Он обычно примерно равен ширине и показывает, на сколько элементов нужно сдвинуться в памяти, чтобы перейти на следующую строчку изображения. Мы мыслим об изображении как о матрице, но в памяти это линия. Отступ нужен, чтобы сделать из линии прямоугольник.

При ознакомлении со списком параметров у вас мог возникнуть вопрос: зачем передавать w и h , если kernel знает глобальный размер? Ответ прост: глобальный размер иногда больше реального. Так делают, чтобы он был кратен размеру рабочей группы. Передача настоящего размера параметром позволяет избежать неправомерного доступа к памяти.

Host

выбираем устройство, на котором будем запускать код;
компилируем kernel под это устройство;
создаём всё, что нужно для запуска, и вызываем kernel-код;
сохраняем изображение, которое он построил, и смотрим, что получилось.

Программа, использующая OpenCL, выглядит так:

Ошибки в этом примере обрабатывать не будем

Программа начинается с include-директивы. Внутри main инициализируем OpenCL вызовом ocl_init . Жёстко зададим размер изображения, которое у нас получится: 1200 x 640 пикселей.

Чтобы main не разрастался, напишем отдельную функцию для выбора устройства.

В этом примере не будем умничать и возьмём первое же устройство. Вначале попробуем найти GPU. Если не получилось — ищем CPU. Да, OpenCL может выполняться и на обычном процессоре, если установлен соответствующий драйвер. В реальной программе можно получить список всех устройств и предоставить возможность выбора пользователю.

Проверять железо на поддержку не нужно, потому что библиотека OpenCL выдаёт только те девайсы, на которых она будет работать.

В этом примере синим цветом выделены функции из API OpenCL. Здесь также использованы три типа данных OpenCL:

cl_platform_id — задаёт платформу, то есть группу устройств;
cl_device_id — задаёт устройство;
cl_int — 32-битное знаковое число.

// main cl_device_id device = create_device(); // Функции clCreateContext из API OpenCL передаём список устройств. // В большинстве случаев — одно устройство. cl_context context = clCreateContext(NULL, 1, err);

Теперь всё готово к тому, чтобы переходить к запуску kernel-кода. Но для начала нужно его скомпилировать. Сделаем это такой функцией:

Скомпилированный код программы можно сохранить и переиспользовать

Функция сперва загружает исходный код, а затем выполняет компиляцию. Это простой пример, и в нём не затронуты следующие возможности, необходимые в реальной программе:

передача параметров компилятору, например значений макросов;
получение лога компиляции — ошибок и предупреждений;
сохранение бинарного кода для последующего использования на том же компьютере;
обработка ошибок.

Код можно прочитать из файла или встроить в программу как ресурс

Далее нужно создать объекты и ресурсы, необходимые для запуска. Мы взяли kernel-код, но не сказали, какая функция из него нам нужна. В нём может быть несколько точек входа, kernel-функций. А нам нужна та, которую мы назвали draw_mandelbrot . Пишем её название как строковый литерал. Функция clCreateKernel получает нужную точку входа по имени функции. Проведём эту и другие манипуляции в main :

Ещё одно важное понятие — очередь. Она нужна для взаимодействия программы и устройства. Вы даёте задачи для GPU, они поступают в очередь. Так как всё происходит асинхронно, видеокарта может быть уже занята задачей, которую ранее поставило ваше приложение либо другая программа, работающая на том же компьютере. Новая команда встанет в очередь и выполнится, когда устройство будет готово. В рамках одной очереди команды выполняются последовательно, но и это можно настроить параметрами функции clCreateCommandQueue .

Также понадобится буфер для записи результата. Мы получили его, выделив нужное количество видеопамяти функцией clCreateBuffer . Параметр CL_MEM_WRITE_ONLY говорит о том, что kernel-код не будет читать из этой памяти. Это позволит применить дополнительные оптимизации.

Теперь всё готово. Осталось только запустить kernel. Размер рабочей группы выберем стандартный: 256 в ширину и 1 в высоту. Глобальный размер задаётся параметрами w и h . Но нужно позаботиться о том, чтобы он делился на локальный. Для этого воспользуемся функцией align , которую я приведу позже:

Вызовом clSetKernelArg задаются все аргументы, которые есть у kernel-функции draw_mandelbrot. Они сохраняются внутри объекта cl_kernel. Особо можно отметить, что clSetKernelArg — это единственная непотокобезопасная функция OpenCL. Все остальные вызовы можно делать из разных тредов вашей программы без каких-либо синхронизаций. Разумеется, если вы уверены в целостности своих данных.

Полный код задания аргументов:

err |= clSetKernelArg(kernel, 0, sizeof(float), err |= clSetKernelArg(kernel, 1, sizeof(float), err |= clSetKernelArg(kernel, 2, sizeof(float), err |= clSetKernelArg(kernel, 3, sizeof(float), err |= clSetKernelArg(kernel, 4, sizeof(cl_int), err |= clSetKernelArg(kernel, 5, sizeof(cl_int), err |= clSetKernelArg(kernel, 6, sizeof(cl_mem), err |= clSetKernelArg(kernel, 7, sizeof(cl_int),

clEnqueueNDRangeKernel — функция, отдающая команду видеокарте на запуск kernel. Она поместит задачу в очередь.

Kernel запишет результат в видеопамять, заданную буфером buff. В этот буфер будет записано изображение множества Мандельброта. Затем мы запросим чтение результата в обычную оперативную память. За трансфер из GRAM в RAM отвечает функция clEnqueueReadBuffer .

Не хватает только функции align :

Чтобы глобальный размер был кратен размеру рабочей группы, воспользовались такой функцией

В конце мы ждём, пока в очереди выполнятся все команды, в том числе и чтение. За это отвечает clFinish . Вопреки названию, она не завершает очередь, а ждёт завершения всех задач, которые в неё поставлены.

Подберём параметры и запустим функцию, читая результат в вектор:

Осталось совсем немного — сохранить результат. Для этого выберем самый простой формат изображений — PPM. Вот функция для сохранения в него:

Сборка и запуск

Итак, программа написана. Но сохраним интригу и прежде чем показать результат её работы, разберём сборку и запуск. Многие программисты на C++ сталкивались с проблемами сборки: то что-то не компилируется, то возникают конфликты, то странные ошибки. С OpenCL в этом плане не всё так плохо. Главная причина: библиотека OpenCL — динамическая. Так задумано её использование.

Для обращения к ней нужен лишь несложный интерфейс, который умещается в нескольких файлах библиотеки clew. Clew — это маленькая библиотечка, позволяющая делать замечательные вещи. Она легко собирается, сама ищет динамическую библиотеку OpenCL и содержит все необходимые include-файлы.

Библиотека clew — не единственный поставщик cl.h. Альтернатива — тяжеловесный инструмент CUDA Toolkit или готовый пакет на системах с пакетным менеджером. При компоновке нужно добавить соответствующую библиотеку. Clew состоит всего лишь из одного компилируемого файла, и можно просто добавить его в проект.

Для запуска программы нужно, чтобы она увидела динамическую библиотеку OpenCL, а библиотека OpenCL должна увидеть видеокарту. На Windows для работы с видеокартой достаточно драйвера. Для других систем или для использования CPU, вероятно, придётся устанавливать специальные драйверы. Вот ссылка на драйвер для интеловских процессоров.

Как правило, самый универсальный способ установить всё и сразу — установить CUDA Toolkit. В него входит cl.h, динамическая библиотека OpenCL и драйвер карты Nvidia.

Ну что ж, программа готова, можно запускать!

Изображение множества Мандельброта, полученное написанной программой

Итоги

как передавать массивы в kernel;
как и для чего использовать локальную память;
как синхронизировать work-item’ы между собой;
как пользоваться векторными типами, например int4 , float8 ;
для чего нужны события;
как пользоваться текстурами.

Напоследок я приготовил схему, которая напомнит, как по шагам написать host-код для OpenCL. Думаю, начинающим она будет полезна.

В этой статье мы подробно разобрали написание простейшей программы для OpenCL. В следующей части рассмотрим алгоритмы на GPU и напишем более сложный kernel-код.

Полный код примера опубликован в репозитории на GitHub.

Источник: temofeev.ru

What is OpenCL?

OpenCL (trademark of Apple Computers Inc.) is an open, royalty-free industry standard that makes much faster computations possible. The standard is controlled by non-profit standards organisation Khronos . By using this technique and graphics cards (GPUs) or extensions of modern processors you can for example convert a video in 20 minutes instead of 2 hours.

Programming the GPU was a very difficult task done by specialised teams and universities, but since 2010 it is in reach of more companies.

Below is a video which explains the differences between single-core, multiple core (starting at 1:27) and OpenCL (starting at 2:32).

You can read more about the engineering ins and outs of the standard at http://www.khronos.org/opencl/ .

How OpenCL works

OpenCL is an extension to existing languages. It makes it possible to specify a piece of code that is executed multiple times independently from each other. This code can run on various processors – not only the main one. Also there is an extension for vectors (float2, short4, int8, long16, etc), because modern processors have support for that.

So for example you need to calculate Sin(x) of a large array of one million numbers. OpenCL detects which devices could compute this for you and gives some statistics of each device. You can pick the best device, or even several devices, and send the data to the device(s). Normally you would loop over the million numbers, but now you say something like: “Get me Sin(x) of each x in array A”. When finished, you take the data back from the device(s) and you are finished.

As the compute-devices can do more in parallel and OpenCL is better in describing independent functions, the total execution time is much lower than conventional methods.

5 questions on OpenCL

Q: Why is it so fast?
A: Because a lot of extra hands make less work, the hundreds of little processors on a graphics card being the extra hands. But cooperation with the main processor keeps being important to achieve maximum output.

Q: Does it work on any type of hardware?
A: As it is an open standard, it can work on any type of hardware that targets parallel execution. This can be a CPU, GPU, DSP or FPGA.

Q: How does it compare to OpenMP/MPI?
A: Where OpenMP and MPI try to split loops over threads/servers and is CPU-oriented, OpenCL focuses on getting threads being data-position aware and making use of processor-capabilities. There are several efforts to combine the two worlds.

Q: Does it replace C or C++?
A: No, it is an extension which integrates well with C, C++, Python, Java and more.

Q: How stable/mature is OpenCL?
A: Currently we have reached version 1.2 and is 3 years old. OpenCL has many predecessors and therefore quite older than 3 years.

Want to know more? Get in contact!

We are the acknowledged experts in OpenCL, CUDA and performance optimization for CPUs and GPUs. We proudly boast a portfolio of satisfied customers worldwide, and can also help you build high performance software. E-mail us today

Источник: streamhpc.com

Вычисляем на видеокартах. Технология OpenCL. Часть 1b. Пишем для OpenCL

В предыдущих сериях

Kernel

Host

Сборка и запуск

Итоги

What is OpenCL?

How OpenCL works

5 questions on OpenCL

Want to know more? Get in contact!

Для чего нужна программа компилятор

Программа смешанное белье для чего

Программа аршин метрология для чего

Программа чек пфр для чего

Токси риск программа для чего

Для чего предназначена программа стили

Для чего нужна программа тренажер

Для чего нужна программа apkpure