Application of NVIDIA CUDA parallel programming technology in the task of melting a spherical particle

Sechenov Pavel PhD in Technical Science Docent, the department of Applied Information Technologies and Programming, Siberian State Industrial University 654007, Russia, Kemerovskaya oblast', g. Novokuznetsk, ul. Kirova, 42 pavesa89@mail.ru Other publications by this author

Olennikov Aleksei Aleksandrovich PhD in Technical Science Associate Professor, Department of Heat, Gas and Water Supply, Water Drainage and Ventilation, Siberian State Industrial University 654007, Russia, Kemerovskaya oblast', g. Novokuznetsk, ul. Kirova, 42 tgsv-sibsiu@mail.ru Other publications by this author

Целью работы является сравнение производительности программы расплавления шарообразной частицы решаемой на центральном процессоре (CPU) на языке C# и с применением технологии параллельного программирования CUDA на графическом процессоре (GPU) на языке CUDA C.

Впервые программно-аппаратная архитектура CUDA (Compute Unified Device Architecture) появилась в феврале 2007 года, представив программистам использовать технологию GPGPU (General-Purprose computing on Graphics Processing Units – неспециализированные вычисления на графической карте), благодаря которой на привычных языках высокого уровня (прежде всего – Си) можно реализовывать алгоритмы, которые выполняются на графических ускорителях GeForce восьмого поколения и старше ^[1].

Одной из важнейших характеристик любого вычислительного устройство является производительность. Процессоры архитектуры x86, c момента появления в 1978 году, увеличили свою тактовую частоту с 4,77 МГц до 3 ГГц, т.е. более чем в 600 раз, однако в последние несколько лет рост частоты более не наблюдается ^[1]. Это связанно как с ограничениями технологии производства микросхем, так и с тем, что энергопотребление (а значит и выделение тепла) пропорционально четвертой степени частоты. Таким образом, увеличение тактовой частоты всего в 2 раза приводит к увеличению тепловыделения в 16 раз ^[2]. В настоящее время рост производительности идёт в основном за счет увеличения числа параллельно работающих ядер, т.е. за счет параллелизма (даже в современных смартфонах число ядер 4-8, а тактовая частота ниже чем у персонального компьютера, что ограничивается возможностями отвода тепла в смартфонах).

Максимальное ускорение, которое можно получить от распараллеливания программы на N процессоров (ядер), дает закон Амдала (Amdahl Law)^[2]:

`S=1/((1+P)+P/N)` (1)

где P – это часть времени выполнения программы, которая может быть распараллелена на N процессоров.

За последние 10 лет появилось множество обучающей литературы по технологии CUDA отечественного ^[1-3] и зарубежного производства ^[4-5]. В связи с тем, что расчеты на графическом процессора выполняются быстрее, чем на центральном процессоре, возрастает интерес к моделированию систем с большим количеством частиц ^{[6, 7]}.

На сайте производителя графических процессоров NVIDIA обучающий онлайн курс ^[8] позволяет освоить основы параллельного программирования. Автор данного курса производит измерение скорости программ на CPU и GPU как с использованием низкоуровневого программирования на языке CUDA C, так и с использованием программного стандарта параллельного программирования OPENACC, который позволяет программисту абстрагироваться от особенностей инициализации графического процессора, вопросов передачи данных на сопроцессор и обратно. При этом наибольшую производительность для параллельных алгоритмов дает программирование на низкоуровневом языке CUDA C в сравнении с использованием OPENACC.

Для увеличения производительности алгоритма требуется выполнение следующих условий:

- алгоритм программы должен распараллеливаться (при этом рост производительности можно рассчитать по закону Амдала);

- наличие графического процессора с технологией CUDA;

- наличие среды разработки (в данном случае это Visual Studio 2013);

- наличие компилятора языка CUDA C (nvcc, который входит в состав CUDA SDK);

- знание языков С#, CUDA C.

Задача расплавления частицы железной руды ^[9], которая является одной из задач моделирования процессов в колонном струйно-эмульсионном реакторе ^[10-12], хорошо распараллеливается и состоит в расчете скорости расплавления одного слоя. В основу растворения кусков принята следующая гипотеза: считается, что куски представляются в виде шара, а в шаре рассматриваются достаточно тонкие последовательно смываемые слои. В слоях рассматривается баланс энтальпии, необходимый для прогрева и усвоения (смыва) слоя, и поток тепла из окружающий среды, необходимый для обеспечения величины этой энтальпии. Для слоя шара объёмом dV за время поступает количество теплоты ^[13]:

`dQ=pc_(p)T_(pl)dV` (2)

где: `p` – плотность; `c_(p)` – теплоёмкость; `T_(pl)` – температура плавления.

С другой стороны, за время через площадь поверхности S данного слоя шара поступает из окружающей среды с температурой количество теплоты ^[13], в соответствии с соотношением (3):

`dQ=alpha(T_(sr)-T_(sl))Sd tau` (3)

где: `alpha` – коэффициент теплоотдачи, `T_(sr)` – температура среды; `T_(sl)` – температура слоя.

Из (2) и (3) уравнений получаем:

`pc_(p)T_(pl)dV= alpha(T_(sr)-T_(sl))Sd tau` (4)

Данные по теплофизическим константам `alpha,` `p` , `c_(p)` , `T_(pl)` взяты из источников ^[14],[15].

Отсюда находится время растворения слоя по формуле (5):

`tau=(pc_(p)T_(pl)dV_(sl))/(alpha(T_(sr)-T_(sl))S)` (5)

Передача тепла в шаре, происходит по следующей схеме: пока внешний слой расплавляется, следующий за ним слой нагревается, как показано на рисунке 1.

Рисунок 1 – Передача тепла на i и i+1 шаге

Перейдем к рассмотрению особенности программной реализации на языке CUDA C. При программировании на GPU используются два понятия: хост, который связан с центральным процессором (host – CPU) и девайс, который связан с графическим процессором (device – GPU). Вызов параллельных частей происходит в функции ядра:

Kernel <<<nBlk, nTid>>> (args),

где: Kernel – название функции ядра;

nBlk – количество блоков функции;

nTid – количество нитей функции;

args – аргументы функции.

В ходе программы на GPU каждая нить имеет свой идентификационный номер:

thredID = threadIdx.x + blockIdx.x * blcockDim.x

где: threadIdx – номер нити в блоке;

blockIdx – номер блока, в котором находится нить;

blockDim – размер блока.

Программа состоит из двух функции: основной – Main() выполняем на CPU и функции расчета MyFunc() выполняемой на GPU.

На рисунке 2 показан алгоритм реализации функции на GPU, который состоит из следующих блоков:

1) вызов функции на GPU с параметрами (массив расчетов, количество расчетов, начальный радиус, шаг, плотность, температура плавления, коэффициент теплоотдачи, температура среды);

2) определение идентификатора нити;

3) определение начального и конечного радиуса;

4) определение времени плавления для каждого слоя, которое рассчитывается параллельно.

Рисунок 2 – Алгоритм реализации функции на GPU

На рисунке 3 показан алгоритм реализации функции на CPU, который состоит из следующих стадий:

1) вызов главной функции;

2) выделение памяти для хоста и девайса;

3) определение количества блоков в потоке;

4) копирование данных с хоста (CPU) на девайс (GPU);

5) вызов функции ядра myFunc с требуемыми параметрами;

6) копирование данных с девайса (GPU) на хост (CPU);

7) суммирование времен расплавления слоев;

8) освобождение памяти хоста и девайса.

Рисунок 3 – Алгоритм реализации функции на CPU

В таблице 1 представленно время выполнения задачи на CPU и GPU в зависимости от количества слоев выделяемых в частице.

Таблица 1 – Сравнение времени выполнения задачи на CPU и GPU.

Как видно из таблицы 1 выигрыш в производительности GPU составляет от 11.4 до 59.4 раз. Наибольший выигрыш по производительности виден при увеличении количества слоев, так на GPU за одну секунду можно обработать на два порядка больше слоев, чем за тоже время на CPU.

Технология параллельного программирования NVIDIA CUDA позволила увеличить производительность распараллеливаемого алгоритма расплавления частицы сложности N от 11 (при 10 000 слоев) до 60 раз (при 100 000 000). Для использования технологии CUDA требуется наличие графического процессора с поддержкой данной технологии, среда разработки и компилятор языка CUDA, знание языка CUDA C, а также хорошее знание задачи и возможности ее распараллеливания. В данной статье не рассматривается директива OPENACC, которая не требует знания языка CUDA С, но по данным автора ^[8] дает меньший прирост производительности в 5-10 раз.