Реализация вычислений на графическом процессоре с использованием платформы Nvidia CUDA

В общем случае создание системы хранения и обработки данных видеонаблюдения на базе грид-технологий происходит следующим образом.

• Определяются цели, задачи и общие требования к системе видеонаблюдения на базе грид-технологий.

• Выбираются исходные данные для создания системы: информация о параметрах системы видеонаблюдения (количество камер, глубина архива, качество записи и др.), варианты использования камер (уличная камера, внутренняя, сценарии использования и т.д.).

• Исследуется модель системы для определения приемлемых характеристик системы видеонаблюдения на базе грид: сравниваются различные конфигурации систем по разным показателям.

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

• Оценивается система по таким показателям, как надежность (пропускная способность сети, загрузка ресурсов хранения данных) и быстродействие (размеры очередей, время ожидания файлов в очереди, загрузка каналов связи).

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

Разработана система ЕаъуОпй для моделирования видео-грид-систем с целью их оптимизации по заданным требованиям к системе (цена, скорость, надежность, выделенные ресурсы). На базе инфраструктуры ЛИТ ОИЯИ реализована система

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

- показано, что в сравнении с централизованным хранилищем данных, применяемым в системах видеонаблюдения, распределенные средства хранения данных на базе грид-технологий имеют ряд преимуществ;

- предложена методика создания системы хранения и обработки данных видеонаблюдения на базе грид-технологий;

- предложен программный инструментарий, позволяющий определить работоспособность системы видеонаблюдения на базе грид-технологий;

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

Авторы благодарны ученым и специалистам Международного университета природы, общества и человека «Дубна»: профессору В.В. Коренькову за постановку задачи и поддержку, Ю.А. Крюкову за помощь в получении экспериментального материала, а также профессорам Е.Н. Черемисиной и С.В. Ульянову за полезные обсуждения.

Литература

1. Интеллектуальные системы безопасности, видеонаблюдение. URL: http://www.iss.ru/products/calc/ (дата обращения: 22.12.2009).

2. Применение систем хранения данных в системах видеонаблюдения. URL: http://www.bitech.ru/ru/resources/articles/ storage_for_video.html (дата обращения: 22.12.2009).

3. Ашихмин И. Применение технологий сетевого хранения данных для цифровых систем видеонаблюдения // Алгоритм безопасности. № 6, 2007. URL: http://www.timcom-pany.ru/article40.html (дата обращения: 22.12.2009).

4. Нечаевский А.В., Кореньков В.В. Пакеты моделирования DataGrid // Системный анализ в науке и образовании (Электрон. журн.). 2009. Вып. 1. URL:http://sanse.ru (дата обращения: 22.12.2009).

РЕАЛИЗАЦИЯ ВЫЧИСЛЕНИИ НА ГРАФИЧЕСКОМ ПРОЦЕССОРЕ С ИСПОЛЬЗОВАНИЕМ ПЛАТФОРМЫ NVIDIA CUDA

А.С. Белозёров; В.В. Коробицын, к.ф.-м.н.

Продемонстрирован подход к организации вычислений на графическом процессоре с помощью платформы NVIDIA CUDA. В результате вычислительного эксперимента по реализации численного метода решения системы обыкновенных дифференциальных уравнений получены более высокие показатели производительности графического процессора по сравнению с центральным процессором.

Ключевые слова: высокопроизводительные вычисления, вычисления на графическом процессоре, методы Рунге-Кутты.

Системы обыкновенных дифференциальных уравнений используются для описания большого

класса математических моделей. Анализ таких моделей зачастую производится численно. С рос-

том количества уравнений в системе увеличиваются вычислительные затраты на построение траекторий решений. Для построения фазовых портретов и анализа бифуркаций необходимо многократно вычислять решения системы при разных значениях параметров и начальных данных системы. Возникает потребность в быстром вычислении большого количества траекторий. Такая вычислительная задача хорошо поддается распараллеливанию. Однако существует несколько принципиально разных подходов к организации параллельных вычислений: на центральном процессоре в многопоточном режиме, на параллельном компьютере с общей памятью, на вычислительном кластере, на графическом процессоре. Последний вариант реализации представляется наиболее интересным.

Производительность современных графических процессоров (GPU) в вычислительных задачах превосходит производительность современных центральных процессоров (CPU), но при этом их стоимость сопоставима. Это преимущество используется для решения различных вычислительных задач в рамках направления GPGPU - задачи общего назначения для GPU (см. http://www.gp-gpu.ru/). С помощью графических библиотек OpenGL и DirectX создаются программы-шейдеры для обеспечения вычислений на графическом процессоре. Однако данный путь предполагает определение задачи в терминах вершин, примитивов и текстур, принятых для GPU, с чем связаны накладные расходы, а также трудности с доступом к памяти из пиксельных, вершинных и геометрических программ.

Платформа NVIDIA CUDA скрывает многие недостатки подхода непосредственного построения программ-шейдеров, предоставляя разработчику прямой доступ к ресурсам GPU, удобный программный интерфейс для параллельных вычислений и компилятор для языка C.

Данная статья посвящена изучению способа реализации численного метода решения систем обыкновенных дифференциальных уравнений на платформе NVIDIA CUDA.

Особенности реализации алгоритмов на платформе NVIDIA CUDA

При программировании на CUDA используется следующая модель вычислений. С точки зрения программиста, GPU является сопроцессором для CPU, имеющим собственную память и способным параллельно выполнять огромное количество потоков. Параллельные участки выполняются на GPU в виде ядер (kernel), запускающихся одновременно в большом количестве потоков. Потоки (threads) группируются в блоки (blocks), которые объединяются в сетку (grid). Потоки внутри одного блока могут взаимодействовать между собой.

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

Для компиляции CUDA-приложений используется компилятор NVCC [1]. Расширения для языка C подразделяются на четыре группы:

• спецификаторы типов функций, указывающие, где будет выполняться функция (на CPU или GPU) и откуда она может быть вызвана;

• спецификаторы типов переменных для уточнения, в какой области памяти (constant, device, global) будет располагаться переменная;

• директива для уточнения, каким образом ядро должно запуститься на GPU;

• встроенные переменные для обозначения размеров сетки и блоков, индексов потоков внутри сетки и блока.

Запуск потока на выполнение и его конфигурацию опишем следующим образом: _global_ void KernelFunc(...); dim3 DimGrid(100, 50); // размер сетки dim3 DimBlock(4, 8, 8); // размер блока KernelFunc<<<DimGrid, DimBlock>>>(...);

// Запустить ядро

Для выделения и освобождения памяти на GPU используют функции cudaMallocO и cudaFree(); для выделения и освобождения на CPU - cudaMallocHost() и cudaFreeHost(). Копирование данных осуществляется с помощью функции cudaMemcpy().

Приведем пример программы на CUDA, осуществляющей вычисление простейшего арифметического выражения на графическом процессоре:

int num=5;

// Ядро, выполняемое на GPU _global_ void myKernelFunc (int* a) { int c = a * 2; }

// Программа, выполняемая на CPU int main

{ CUT_DEVICE_INIT(); // Инициализация int* A,devA; A=#

// Выделение памяти cudaMalloc((void**)&devA, sizeof(int)); cudaMemcpy(devA,(void*)(&A),sizeof(int), cudaMemcpyHostToDevice);

// Запуск ядра

myKernelFunc<<< 1,dim3(300,1)>>> (devA);

cudaFree(devA);

return 0; }

Сравнение производительности GPU и CPU

Оценка производительности вычислений на GPU и CPU производилась на известной системе уравнений Лотки-Вольтерра, так называемой модели «хищник-жертва»: dx „ 2

—=ttix-Pxy-YiX ,

dy П 2 —=a2y+Pxy-Y2У • dt

В эксперименте в каждом потоке производится вычисление траектории системы уравнений, имеющей одинаковые параметры, но отличающейся начальными данными:

xk =10+2k, yk = 8+k, tk = 0, k k

здесь x0 , y0 - начальная точка траектории решения в момент tk0 потока с номером k. Используемые значения параметров системы: a1 = 0.1, р=0.1, у 1=0,

a2 =-1, Y2 = 0.

Условия остановки процесса решения одинаковы для всех потоков: достигнуто заданное количество шагов N. Локальная точность решения выбрана невысокой, е=0.001, поскольку CUDA поддерживает только 32-битный формат чисел с плавающей точкой. В алгоритме использована вложенная схема Мерсона, обеспечивающая оценку локальной погрешности решения, и процедура автоматического выбора шага интегрирования.

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

В таблице приведены результаты вычислительного эксперимента для сравнения производительности двух реализаций.

Была исследована эффективность на длинных (10000 шагов) и коротких (500 шагов) решениях. Реализация метода Мерсона на CUDA демонстрирует очень высокую производительность по сравнению с CPU и отличную масштабируемость при увеличении количества траекторий.

При количестве траекторий более 256 реализация на CUDA при использовании видеокарты на базе 8800GT показывает 10-15-кратное преимущество перед реализацией на процессоре Intel Core 2 Duo E8200.

Учитывая, что стоимость видеокарты и процессора примерно одинакова, покупка видеокарты для подобных расчетов является намного более выгодным приобретением.

Сравнение показателей производительности реализаций

№ Коли- Количе- Производитель- Производи-

экс- чест- ство ша- ность, шагов/сек. тельность,

пе- во гов N в CPU Core 2 Duo шагов/сек.

ри- траек- одной E8200 2.66Ghz CUDA на

мента торий траекто- (один поток), GPU GeForce

рии 4096Mb RAM 8800GT,

512Mb RAM

1 1 10000 71118 213354

2 2 10000 85341 320032

3 3 10000 87281 480048

4 4 10000 94824 512051

5 5 10000 94127 640064

6 6 10000 93667 480048

7 10 10000 95531 640064

8 50 10000 70401 627513

9 100 10000 52728 524642

10 256 10000 45476 807174

11 512 10000 43511 799299

12 1024 10000 49741 Out of memory

13 10 500 80160 106880

14 50 500 89066 229028

15 100 500 91611 400800

16 256 500 76713 390875

17 512 500 77804 373108

18 1024 500 83122 377396

Проведенный эксперимент показал, что платформа NVIDIA CUDA является удачным выбором для выполнения распараллеливаемых вычислительных экспериментов, если не предъявляются высокие требования к точности результата.

К недостаткам платформы можно отнести поддержку только 32-битных чисел с плавающей точкой, невозможность динамического выделения памяти на GPU, отсутствие поддержки рекурсии и указателей на функции, что ограничивает возможности приложений, реализуемых на CUDA.

Возможности платформы были продемонстрированы при параллельном решении систем дифференциальных уравнений методом Мерсона. Реализация на CUDA показала значительный выигрыш в производительности по сравнению с тем же алгоритмом, выполняемым на центральном процессоре - до 10 раз при решении уравнений с большим количеством необходимых итераций.

Литература

1. NVIDIA CUDA Programming Guide 1.1 URL: http://www.nvidia.com/object/cuda_develop.html (дата обращения: 01.04.2009).

2. Хайрер Э., Нёрсетт С., Ваннер Г. Решение обыкновенных дифференциальных уравнений. Нежесткие задачи. М.: Мир, 1990. 512 с.

3. Pharr M., Randima F. GPU Gems 2: Programming Techniques for High-Performance Graphics and General-Purpose Computation. Addison-Wesley, 2005. 812 pp.