Оценка производительности суперкомпьютерного центра «Политехнический» на примере задачи решения системы линейных уравнений. Отчётпо курсовой работе по дисциплине Высокоскоростные сетевые технологии суперкомпьютеровТема работы Оценка производительности суперкомпьютерного центра Политехнический на примере задачи решения системы линейных уравненийВыполнилаО.
 Скачать 278.81 Kb.
|
|
САНКТ-ПЕТЕРБУРГСКИЙ ПОЛИТЕХНИЧЕСКИЙ УНИВЕРСИТЕТ ПЕТРА ВЕЛИКОГО Институт прикладной математики и механики Высшая школа прикладной математики и вычислительной физики ОТЧЁТ по курсовой работе по дисциплине «Высокоскоростные сетевые технологии суперкомпьютеров» Тема работы: Оценка производительности суперкомпьютерного центра «Политехнический» на примере задачи решения системы линейных уравнений Выполнила О. В. Саксина студентка группы 3630201/70101 Руководитель М. А. Курочкин Санкт-Петербург, 2021 г. Содержание Введение 2 1. Постановка задачи 3 2. Аппаратно-программная платформа Nvidia CUDA 4 2.1 Потоковая модель 4 2.2 Устройство памяти 4 3. Суперкомпьютерный центр “Политехнический” 6 3.1 Состав 6 3.2 Характеристики 6 3.3 Технология подключения 6 4. Постановка решаемой задачи 9 5. Алгоритм решения задачи 9 6. Описание эксперимента 10 6.1 Проведение измерений 10 6.2 Анализ результата 12 Заключение 13 Список литературы 14 1 Введение GPGPU (General-Purpose Graphics Processing Units) – техника использования графического процессора видеокарты для общих вычислений, которые обычно проводит центральный процессор, т.е. это набор аппаратных и программных технологий, позволяющие использовать графические процессоры, для ускорения многих не графических приложений. Graphics Processing Units (GPUs) являются высокоэффективными многоядерными процессорами, способными к сложным вычислениям и очень высокой пропускной способности данных. Сегодня существует несколько технологий, реализующих возможность программирования графических процессоров при помощи C-подобных языков программирования и обеспечивающих переносимость кода. CUDA (Compute Unified Device Architecture) – программно-аппаратная архитектура, позволяющая производить вычисления с использованием графических процессоров NVIDIA, поддерживающих технологию GPGPU. Программная часть технологии CUDA - это надстройка языка С, в котором присутствуют дополнительные ключевые слова и синтаксис, позволяющий реализовывать код для его последующего выполнения на GPU устройстве. OpenCL (Open Computing Language) это спецификация, описывающая технологию параллельного программирования, которая в первую очередь ориентирована на GPGPU. Для развития спецификаций OpenCL был образована группа разработчиков Khronos Compute, в неё вошли nVidia, AMD, IBM, Intel, ARM и др. В отличии от nVidia CUDA в OpenCL изначально закладывалась мультиплатформенность, т.е. OpenCL программа должна без изменений в коде работать на GPU разных типов (разных производителей). Такая программа без изменений должна работать даже на CPU без GPU, хотя в этом случае она может выполняться существенно медленнее чем на GPU. Тесты LINPACK служат для измерения вычислительной производительности компьютеров при обработке чисел с плавающей запятой. Созданы Джеком Донгаррой в 1979, измеряют скорость решения компьютером плотной системы линейных уравнений (СЛАУ) Ax=b, где A является матрицей размера n на n. Подобная задача часто возникает в области машиностроения. Последняя версия этих тестов производительности используется для составления рейтинга TOP500, списка, в котором перечислены самые высокопроизводительные суперкомпьютеры в мире. Целью создания тестов является оценка скорости решения компьютером реальных задач. Однако оценка при помощи одного теста является упрощением, поскольку ни одна вычислительная задача не может отражать общую производительность компьютерной системы. Производительность теста LINPACK предоставить данные для уточнения пиковой производительности, предоставляемой производителем 2 компьютера. Пиковая производительность — это максимальная теоретическая производительность, которую может достичь компьютер, рассчитанная как произведение тактовой частоты процессора на число операций выполняемых за такт. Фактическая производительность всегда будет ниже максимальной производительности. Производительность компьютера представляет собой сложную характеристику, которая зависит от множества взаимосвязанных компонентов. Производительность, измеренная эталонным тестом LINPACK показывает количество операций над 64-битными числами с плавающей запятой (сложений и умножений), которые компьютер выполнял за секунду, соотношение, обозначаемое «FLOPS». Однако производительность компьютера при работе с реальными приложениями, вероятно, будет значительно ниже максимальной производительности, достигаемой при выполнении соразмерного теста LINPACK. 3 1. Постановка задачи Данная работа направлена на изучение платформы OpenCL, реализации соб- ственного алгоритма решения системы линейных уравнений и исследование скорости выполнения решения поставленной практической задачи решения системы линейных уравнений. Необходимо: ● Изучить аппаратно-программную платформу OpenCL ● Реализовать алгоритм решения системы линейных уравнений методом Гаусса с помощью OpenCL ● Сравнить время исполнения собственной реализации и результаты бенчмарка Linpack ● Провести анализ полученных результатов 4 2 OpenCL OpenCL (Open Computing Language, открытый язык вычислений) — это открытый стандарт для параллельного программирования, предлагающий эффективный и переносимый способ использования возможностей разнородных вычислительных многоядерных платформ (CPU,GPU и др.). Он включает в себя программный интерфейс (API) для координирования параллельных вычислений в среде разнородных процессоров и кроссплатформенный язык, используемый в определённом вычислительном окружении. На конкретной системной платформе стандарт реализуется в виде исполняемых модулей (библиотек), достаточных для запуска пользовательских OpenCL-приложений. Разработка подобных приложений предполагает наличие также C/C++-компилятора системной платформы и набора необходимых заголовочных файлов. 2.1 Основные понятия OpenCL-приложение - это совокупность программного кода, исполняемого на хосте (host) и OpenCL-устройствах (device). Под хостом обычно понимается центральный процессор(CPU) вычислительного устройства (компьютера, планшета, телефона и т.п.), а устройства OpenCL (devices) — некоторый набор вычислительных единиц, соответствующий графическому процессору (GPU), многоядерному CPU или другим процессорам с параллельной архитектурой, доступным из хост-программы. Исполнение OpenCL-приложения, таким образом, включает исполнение хост-программы (на хосте) и исполнение специального кода на одном или нескольких OpenCL-устройствах под управлением хост-программы. Программы, исполняемые на OpenCL-устройствах, содержат одно или несколько ядер (kernels— функции, помеченные специальным ключевым словом __kernel и являющиеся «точкой входа» в исполняемый на устройствах код), при необходимости — вспомогательные функции, вызываемые в тексте ядер, а также, возможно, константные данные. Исполняемые на OpenCL-устройствах ядра пишутся на языке OpenCL C, который основывается на спецификации языка C 1999 года (т.н. C99) со специфическими расширениями и ограничениями. Программная модель OpenCL-приложения —SIMD(Single Instruction Multiple Data) или SIMT(Single Instruction Multiple Threads): код ядра может быть исполнен одновременно на многих вычислительных единицах, каждая из которых работает со «своими» данными. В терминологии OpenCL каждая исполняемая копия кода ядра именуется work-item, она характеризуется своими уникальными индексами: глобальным идентификатором 5 (global ID) и локальным идентификатором (local ID). Этих индексов — два, поскольку локальный индекс характеризует конкретную рабочую единицу в рамках так называемой рабочей группы (work-group), а глобальный индекс уникально характеризует рабочую единицу независимо от принадлежности к какой-либо рабочей группе. 3.2 Потоковая модель Модель вычислений (execution model) описывает абстрактное представление того, как потоки инструкций выполняются в гетерогенной системе. С хостом неразрывно связано понятие хостовой программы (host program) - программного кода, выполняющегося только на хосте. OpenCL не указывает как именно должна работать хостовая программа, а лишь определяет интерфейс взаимодействия с OpenCL-объектами. С точки зрения модели вычислений OpenCL-приложение состоит из хостовой программы и набора ядер (kernels). OpenCL-ядро в общем виде представляет собой функцию, написанную на языке OpenCL C и скомпилированную OpenCL-компилятором. Ядро создается в хостовой программе и затем с помощью специальной команды ставится в очередь на выполнение в одном из OpenCL-устройств. Во время выполнения упомянутой команды OpenCL Runtime System создает целочисленное пространство индексов (integer index space), каждый элемент которого носит название глобальным идентификатором (global ID). Каждый экземпляр ядра выполняется отдельно для каждого значения глобального идентификатора. Экземпляр ядра носит название work-item. Таким образом, каждый work-item однозначно определяется своим глобальным идентификатором. Множество всех work-item разбивается на группы. Такая группа носит название work-group. С каждой work-group сопоставляется свой уникальный идентификатор (work-group ID). Все work-item в одной work-group идентифицируются уникальным в пределах своей группы номером: local ID. Таким образом каждый work-item определяется как по уникальному global ID так и по комбинации work-group ID и local ID внутри своей группы. Все work-item в пределах одной work-group выполняются параллельно на обрабатывающих элементах одного вычислительного модуля OpenCL-устройства. Это гарантируется стандартом, в то время как совершенно не гарантируется, что несколько work-item из разных групп будут выполнены параллельно. Об этом важном свойстве параллелизма необходимо всегда помнить при разработке OpenCL-программ. Пространство индексов N-размерно и обычно носит название NDRange. В случае версии стандарта OpenCL 1.1 размерность N принимает значения 1,2 или 3. Таким 6 образом, сетки координат global ID и local ID N-размерны, т.е. определяются N координатами. Другим важным понятием модели вычислений является контекст, определение которого (при помощи вызова специальных функций OpenCL API) является первой задачей OpenCL-приложения. Контекст определяет среду выполнения ядер, в которую входят следующие компоненты: устройства, сами ядра, программные объекты (program objects, исходный и выполняемый код будущих ядер), объекты памяти (memory objects). Взаимодействие между хостом и OpenCL-устройством происходит посредством команд, помещенных в командную очередь (command-queue). Данные команды ожидают в командной очереди своего выполнения на OpenCL-устройстве. Командная очередь создается хостом и сопоставляется одному OpenCL-устройству после того, как будет определен контекст. Команды делятся на те, что отвечают за: выполнение ядер, управление памятью и синхронизацию выполнения команд в очереди. Команды могут выполняться последовательно (in-order execution) или внеочередно (out-of-order execution). Второй вариант организации очередей поддерживается не всеми платформами, о чем необходимо помнить. 2.3 Устройство памяти Модель памяти OpenCL определяет поведение и иерархию памяти, которая может использоваться приложениями OpenCL. Это иерархическое представление памяти является общим для всех реализаций OpenCL, но отдельные поставщики должны определять, как модель памяти OpenCL отображается на конкретное оборудование. Хост-память Память хоста определяется как область системной памяти, которая напрямую (и только) доступна из хост-процессора. Любые данные, необходимые вычислительным ядрам, должны передаваться в глобальную память устройства OpenCL и из нее с помощью OpenCL API. Глобальная память Глобальная память определяется как область памяти устройства, доступная как для хоста OpenCL, так и для устройства. Глобальная память обеспечивает доступ для чтения и записи к главному процессору, а также ко всем вычислительным блокам устройства. Хост отвечает за выделение и отмену выделения буферов в этом пространстве памяти. Между хостом и устройством происходит синхронизация для управления данными, хранящимися в этой памяти. Главный процессор передает данные из области памяти хоста в глобальную область памяти. Затем, как только ядро запускается для обработки данных, хост теряет права доступа к буферу в глобальной памяти. Устройство берет на себя управление и может читать и писать из глобальной памяти до завершения выполнения ядра. По завершении операций, связанных с ядром, 7 устройство возвращает управление буфером глобальной памяти главному процессору. После восстановления управления буфером главный процессор может читать и записывать данные в буфер, передавать данные обратно в память хоста и освобождать буфер. Постоянная глобальная память Постоянная глобальная память определяется как область системной памяти, доступная с доступом для чтения и записи для хоста OpenCL и с доступом только для чтения для устройства OpenCL. Как следует из названия, типичное использование этой памяти - передача постоянных данных, необходимых для вычислений ядра, от хоста к устройству. Локальная память Локальная память - это область памяти, которая является локальной для одной вычислительной единицы. Хост-процессор не видит и не контролирует операции, которые происходят в этом пространстве памяти. Это пространство памяти позволяет выполнять операции чтения и записи всем обрабатывающим элементам с вычислительными блоками. Этот уровень памяти обычно используется для хранения данных, которые должны совместно использоваться несколькими рабочими элементами. Операции в локальной памяти между рабочими элементами неупорядочены, но синхронизация и согласованность могут быть достигнуты с помощью операций барьера и ограждения. Приватная память Приватная память - это область памяти, которая является частной для отдельного рабочего элемента, выполняемого в элементе обработки OpenCL. Как и в случае с локальной памятью, хост-процессор не видит этой области памяти. Это пространство памяти может быть прочитано и записано всеми рабочими элементами, но переменные, определенные в частной памяти одного рабочего элемента, не видны для другого рабочего элемента. 3. Суперкомпьютерный центр “Политехнический” 3.1 Состав В вычислительном центре представлены три суперкомпьютера. ● «Политехник - РСК Торнадо» - кластер с пиковой производительностью 943 Тфлопс; содержит 668 двухпроцессорных узлов (Intel Xeon Е5 2697 v3), из них - 56 узлов имеют два ускорителя вычислений NVIDIA К40; ● «Политехник - РСК ПетаСтрим» - массивно-параллельный компьютер с уль- травысокой многопоточностью; единственная в России система, способная 8 поддержать более 70 тыс. потоков; пиковая производительность 291 Тфлопс; содержит 288 узлов на сопроцессорах Intel Xeon Phi; ● «Политехник - NUMA» - массивно-параллельная система с кеш-когерентной глобально адресуемой памятью объемом более 12 ТБ; содержит 64 узла, 192 процессора; пиковая производительность 30 ТФлопс. Все вычислительные системы работают с общей системой хранения данных (Seagate ClusterStore, 1.1ПБ), имеют единую систему управления и мониторинга. 3.2 Характеристики В качестве вычислительного ресурса выбран кластер «Политехник - РСК Торнадо», в котором для реализации массивных параллельных вычислений используется один узел с ускорителем вычислений «NVIDIA Tesla К40». Ускоритель вычислений «NVIDIA Tesla К40» относится к ускорителям NVIDIA поколения Kepler и обладает рядом заявленных производителем характеристик: ● 12 Гигабайт памяти ● Скорость полосы пропускания памяти 288 Гб/с ● CUDA-ядер 2880 ● Пиковая производительность для вычислений одинарной точности с плавающей точкой 4.29 ТФлопс Ускоритель поддерживает технологию CUDA. 3.3 Технология подключения Для аутентификации подключений к ресурсам СКЦ используются SSh-ключи, представляющие собой пару файлов: закрытый ключ, хранящийся на компьютере пользователя, и открытый ключ, хранящийся на том компьютере, к которому осуществляется подключение. Для работы с кластером была сгенерирована такая пара ключей в приложении PuTTY. Алгоритм генерации ключей: 1. В утилите PuTTY Key Generator задать тип ключа “SSH-2 RSA”, его размер - 2048 бит, и нажать на кнопку “Generate”. 2. Поводить курсором мыши в пустой области окна. 3. Ввести пароль для созданного ключа. 4. Сохранить публичный ключ. 5. Сохранить закрытый ключ. 6. Файл с открытым ключом предоставить в службу регистрации СКЦ. Доступ к вычислительным ресурсам суперкомпьютера производится с машины login1.hpc.spbstu.ru. Для подключения используется терминальный клиент PuTTY. Подключение происходит следующим образом: 9 1. В программе PuTTY в настройках сессии Conection -> SSH -> Auth, задать путь к файлу с закрытым ключом. 2. В Session -> Host Name указать login1.hpc.spbstu.ru 3. Нажать Open. 4. Ввести логин и пароль. После успешного подключения к суперкомпьютеру для пользователя доступны различные модули. Список модулей можно получить командой module avail. Из всех модулей нам необходим «nvidia/cuda-8.0». Чтобы собрать программу, написанную на языке «CUDA С» необходимо написать команду module add nvidia/cuda-8.0. После чего, ранее недоступная команда nvcc в среде станет доступной. Команда nvcc запускает компилятор, который позволяет скомпилировать программу, написанную на языке «CUDA С». Синтаксис компиляции программы с использованием nvcc: nvcc -о t2.out t1.cu Данная строка компилирует исходный текст программы "t1.cu", написанный на «CUDA С», линкует программу и сохраняет её под названием "t2.out". Для запуска программы на суперкомыотере используется специальная система «SLURM», отправляющая задачу на суперкомпьютер. Для запуска программы на одном из трёх доступных кластеров нужно создать следующий вспомогательный файл «run.slurm»: #! /bin /bash #SBATCH —nodes=l #SBATCH —tasks—per—node=l #SBATCH —cpus—per— task=l #SBATCH —p tornado—k40 #SBATCH —J CUDA_TEST_im #SBATCH —o CUDA_TEST_ im-%j.out #SBATCH —e CUDA_TEST_im-%j.err if [ -f /etc/profile.d/modules-basis.sh ]; then source /etc/profile.d/modules-basis.sh fi module purge ./t2.out Данный скрипт говорит о том, что требуется использовать один узел кластера «tornado-k40», в котором будет работать одна задача на одном процессоре. Имя задачи: "CUDA_TEST_im", файл вывода консоли: ”CUDA_TEST_ im-%j.out”, файл вывода ошибок: "CUDA_TEST_im-%j.err". После настройки запускаемой программы и её конфигурации в файле «run.slurm», необходимо отправить скрипт на исполнение. Для этого исполняется команда sbatch run.slurm 10 Проверять состояние задачи можно посредством команды squeue, отображающей текущий статус всех запущенных задач от имени текущего пользователя. 4 Компиляторы для процессора Intel Программы, написанные на C++, должны быть скомпилированы до их запуска. Компилятор - это программа, которая преобразует текст на одном языке в текст на другом языке. Например, компилятор языка Си переводит программу на языке Си в коды процессора, которые затем могут быть непосредственно исполнены. Процесс обработки текстовых файлов с кодом на языке C++, который упрощенно называют "компиляцией", на самом деле, состоит из четырех этапов. 1. Препроцессинг — обработка текстовых файлов утилитой препроцессора, который производит замены текстов согласно правилам языка препроцессора C/C++. После препроцессора, тексты компилируемых файлов, обычно, значительно вырастают в размерах, но теперь в них содержится все, что потребуется компилятору для создания объектного файла. 2. Ассемблирование — процесс превращения текста на языке C++ в текст на языке Ассемблера. Для компиляторов GNU используется синтаксис ассебмлера AT&T. 3. Компилирование — процесс превращения текстов на языке Ассемблера в объектные файлы. Это файлы состоящие из кодов целевого процессора, но в которых еще не проставлены адреса объектов, которые находятся в других объектных файлах или библиотеках. 4. Линковка — процесс объединения объектных файлов проекта и используемых библиотек в единую целевую сущность для целевой платформы. Это может быть исполняемая программа или библиотека статического или динамического типа. Синтаксис команды компиляции имеет вид: компилятор [опции] файлы [библиотеки] (здесь компилятор - команда вызова компилятора) Основные опции: ● -o - создать выходной файл с заданным именем (без опции создается a.out); ● -c - не изготавливать исполнимый модуль (при компиляции подпрограмм); ● -O -O1,-O2,-O3 - задание уровня оптимизации; ● -g - выполнить компиляцию в отладочном режиме; ● файлы - компилируемые файлы; ● библиотеки - подключаемые библиотеки. 11 В квадратных скобках указываются необязательные компоненты команды. Пакет компиляторов GCC В него входят компиляторы: ● gcc - компилятор языка С; ● g++ - компилятор языка С++; Компиляторы GCC оптимизирующие, поддерживающие три уровня оптимизации (опции -O1, -O2, -O3). На разных программах более эффективной может оказаться та или другая опция. В большинстве случаев наиболее приемлемой бывает опция -O2, при этом ускорение программы может достигать 2-3 раз. Типичная команда компиляции: gcc -O2 -o prog prog.c (для языка С) Пакет компиляторов Intel Широко используется пакет компиляторов Intel Compiler, наилучшим образом оптимизированный под платформу x86, являющуюся основной при построении вычислительных кластеров. ● icc - компилятор языка С; ● icpc - компилятор языка С++; Компиляторы также поддерживают три уровня оптимизации (опции -O1, -O2, -O3, задание опции -O соответствует уровню -O2). Сочетание опций -fast -On, задает режим максимального ускорения программы на соответствующем уровне оптимизации. Для отлаженных программ включение оптимизации обязательно. В большинстве случаев ускорение работы программы может достигать 2-3 раз. 12 4. Постановка решаемой задачи В данной работе требуется решить систему линейных уравнений вида Ax = b. Дано: ● Матрица A - содержит вещественные числа в отрезке [-10000, 10000], размерность (nxn) ● Вектор b - содержит вещественные числа в отрезке [-10000, 10000], длины n Надо: ● Найти вектор, состоящий из корней системы уравнения x 1 , ..., x n , являющихся вещественными числами Ограничения: ● Коэффициенты уравнений генерируются случайным образом 5. Алгоритм решения задачи Пусть исходная система выглядит следующим образом: Её можно записать в матричном виде Ax = b, где матрица A называется основной матрицей системы, b — столбцом свободных членов. Тогда, согласно свойству элементарных преобразований над строками, ос- новную матрицу этой системы можно привести к ступенчатому виду (эти же преобразования нужно применять к столбцу свободных членов): где α 1𝑗 1 , ..., α 1𝑗 𝑟 ≠ 0 Алгоритм решения СЛАУ методом Гаусса подразделяется на два этапа: 1. На первом этапе осуществляется так называемый прямой ход, когда путём элементарных преобразований над строками систему приводят к ступенчатой или треугольной форме, либо устанавливают, что система несовместна. Для этого среди 13 элементов первого столбца матрицы выбирают ненулевой, перемещают содержащую его строку в крайнее верхнее положение, делая эту строку первой. Далее ненулевые элементы первого столбца всех нижележащих строк обнуляются путём вычитания из каждой строки первой строки, домноженной на отношение первого элемента этих строк к первому элементу первой строки. После того, как указанные преобразования были совершены, первую строку и первый столбец мысленно вычёркивают и продолжают, пока не останется матрица нулевого размера. Если на какой-то из итераций среди элементов первого столбца не нашёлся ненулевой, то переходят к следующему столбцу и проделывают аналогичную операцию. 2. На втором этапе осуществляется так называемый обратный ход, суть которого заключается в том, чтобы выразить все получившиеся базисные переменные через небазисные и построить фундаментальную систему решений, либо, если все переменные являются базисными, то выразить в численном виде единственное решение системы линейных уравнений. Эта процедура начинается с последнего уравнения, из которого выражают соответствующую базисную переменную (а она там всего одна) и подставляют в предыдущие уравнения, и так далее, поднимаясь по «ступенькам» наверх. Каждой строчке соответствует ровно одна базисная переменная, поэтому на каждом шаге, кроме последнего (самого верхнего), ситуация в точности повторяет случай последней строки. Метод Гаусса требует O(n 3 ) арифметических операций. Можно заметить, что все вычисления сводятся к однотипным вычислительным операциям над строками матрицы коэффициентов системы линейных уравнений. Как результат, в основу параллельной реализации алгоритма Гаусса может быть положен принцип распараллеливания по данным. В качестве базовой подзадачи можно принять все вычисления, связанные с обработкой одной строки матрицы А и соответствующего элемента вектора b. Выделенные базовые подзадачи характеризуются одинаковой вычислительной трудоемкостью и сбалансированным объемом передаваемых данных. В случае, когда размер матрицы, описывающей систему линейных уравнений, оказывается большим, чем число доступных процессоров, базовые подзадачи можно укрупнить, объединив в рамках одной подзадачи несколько строк матрицы. Однако применение последовательной схемы разделения данных для параллельного решения систем линейных уравнений приведет к неравномерной вычислительной нагрузке процессоров - по мере исключения (на прямом ходе) или определения (на обратном ходе) неизвестных в методе Гаусса для все большей части процессоров все необходимые вычисления будут завершены и процессоры окажутся простаивающими. 14 6. Описание эксперимента 6.1 Проведение измерений Цель: Целью данного эксперимента является сравнение времени исполнения программы собственной реализации метода Гаусса решения СЛАУ, использующую технологию OpenCL, с результатами, полученными от выполнения бенчмарка LINPACK. Linpack Изменяемые параметры: Количество уравнений в системе: 1000, 2000, 5000, 10000, 18000. Число потоков: 32, 64, 128, 512 Методы измерения: Для сбора данных было проведено 10 испытаний для каждого эксперимента, после чего было посчитано среднее значение времени выполнения. Время в таблице указано в секундах. Количество уравнений Число потоков 32 64 128 512 1000 0.007 0.005 0.00034 0.00019 2000 0.021 0.015 0.0061 0.0003 5000 0.181 0.095 0.046 0.0034 10000 0.991 0.534 0.248 0.08 18000 3.042 2.016 0.993 0.395 Собственная реализация Изменяемые параметры: Число блоков: 1, 100, 1000, 10000. Число потоков: 1, 10, 100, 1000 Методы измерения: Для измерения времени исполнения программы событие clGetEventProfilingInfo с платформы OpenCL. Для сбора данных было проведено 10 испытаний для каждого эксперимента, после чего было посчитано среднее значение времени выполнения. Время в таблице указано в миллисекундах. 15 N = 1000 Число блоков Число потоков 1 10 100 1000 1 5983.06934 508.91595 55.60777 32.38478 10 1130.16895 118.00394 16.01032 16.00715 100 262.04871 27.92218 27.90741 27.80695 1000 127.81656 127.81616 127.78708 127.81477 N = 2000 Число блоков Количество потоков 1 10 100 1000 1 18530.12342 1694.30291 105.90328 87.99318 10 1783.96082 271.80518 14.93219 14.94214 100 930.14247 83.70251 83.12854 83.94055 1000 161.99039 161.95752 161.92578 161.18203 N = 10000 Число блоков Число потоков 1 10 100 1000 1 192650.6171 3388.60582 764.7582 831.5754 10 14351.88246 2001.31813 157.262665 157.29749 100 3760.28494 694.81004 694.58912 694.17954 1000 969.98251 969.90248 969.79543 969.03983 16 6.2 Анализ результатов Метод Гаусса требует O(n 3 ) арифметических операций, соответственно количество времени на исполнение программы быстро растёт при росте входных данных. Для 1000 уравнений минимальное время в реализации наблюдается при 100 блоках и 10 потоках, т.е. когда 1 поток отвечает за 1 строку массива. Ускорение по сравнению с последовательным выполнением происходит примерно в 500 раз. При увеличении количества блоков выше 100 время работы не изменяется. Это происходит потому, что суммарное количество потоков во всех блоках превышает 1000, т.е. каждой строке матрицы соответствует больше, чем 1 поток. Таким образом получается, что часть потоков простаивает. Реализация превышает время вычислений теста Linpack. На тесте с системой из 10000 уравнений OpenCL начинает приближаться к результатам Linpack. Это можно объяснить тем, что при увеличении количества уравнений в системе также возрастает количество используемых GPU ресурсов для распараллеливания, что дает существенный прирост к скорости решения на OpenCL. Можно в общем сказать, что результаты решения Linpack лучше на малом числе уравнений в системе, около 1000-5000. На тестах с маленьким количеством уравнений использование GPU замедляет исполнение реализации OpenCL, т.к. тратится дополнительное время на выделение памяти. Еще одной причиной замедления исполнения реализации OpenCL являются накладные расходы, возникающие из-за вызова функций инициализирующих контекст, различные переменные, ядра и проводящие проверки, что на запускаемой машине присутствуют устройства, поддерживающие распараллеливание. 17 Заключение В результате выполнения данной работы была изучена аппаратно-программная платформа Nvidia OpenCL. Было произведено ознакомление с ее аппаратной и программной частью, потоковой моделью и моделью памяти, изучены различные типы памяти. Было произведено ознакомление с вычислительными ресурсами СКЦ Политехнический, изучена технология подключения к нему, изучена технология написания параллельных программ с использованием суперкомпьютера и технологии OpenCL. Был изучен бенчмарк Linpack и способ его запуска. На основе изученных материалов был реализован и распараллелен с помощью OpenCL алгоритм поиска решений системы линейных уравнений методом Гаусса. Также было проведено сравнение реализации алгоритма на OpenCL и бенчмарка Linpack. Для проверки зависимости времени исполнения программы от количества уравнений в системе было проведено некоторое количество испытаний, по результатам которых была заполнена таблица со временем исполнения программ. Полученные результаты были проанализированы. Результаты решения Linpack оказались быстрее при малом числе уравнений в системе. Чем больше уравнений — тем ближе результаты собственной реализации приближались к результатам исполнения Linpack. Это объясняется тем, что с увеличением количества уравнений в системе увеличился эффект ускорения реализации от использования GPU. 18 Список литературы [1] Aaftab MunshiBenedict R. GasterTimothy G. MattsonJames FungDan Ginsburg. OpenCL Programming Guide. – Addison-Wesley, 2011. – 648с. [Электронный ресурс]. http://asu-cs.donntu.org/sites/default/files/images/ doc/opencl.programming.guide.pdf [2] OpenCL Parallel Programming Development Cookbook - Birmingham : Packt Publishing, 2013 год, 302 с. 19 20 Приложение Проведение эксперимента с linpack 1) Загрузить linpack в основную директорию (найти его можно на сайте software.intel.com) 2) Скопировать ссылку на архив linpack для ОС Linux и загрузить на суперкомпьютер при помощи команды wget 3) Распаковывать в той же директории архив, используя команду tar − xvzf и название архива 4) Перенести все файлы из директории linpack в основную директорию, и в ней же создать run.slurm файл: #!/bin/bash #SBATCH --nodes=1 #SBATCH -p tornado #SBATCH -t 10-00:00:00 #SBATCH -J lin_test #SBATCH -o lin_test-\%j.out #SBATCH -e lin_test-\%j.err module purge ./runme_xeon64 5) Файл сохраняется и ставится в очередь командой $ sbatch run.slurm. 6) После выполнения теста в основной директории появляются файлы вывода ошибок 7) После успешного выполнения в выводе находится следующее: This is a SAMPLE run script for running a shared-memory version of Intel(R) Distribution for LINPACK* Benchmark. Change it to reflect the correct number of CPUs/threads, problem input files, etc.. *Other names and brands may be claimed as the property of others. Mon Dec 14 21:36:40 2020 Sample data file lininput_xeon64. Current date/time: Mon Dec 14 21:36:41 2020 CPU frequency: 3.585 GHz Number of CPUs: 2 Number of cores: 32 Number of threads: 32 Parameters are set to: Number of tests: 15 Number of equations to solve (problem size) : 1000 2000 5000 10000 15000 18000 20000 22000 25000 26000 27000 30000 35000 40000 45000 Leading dimension of array : 1000 2000 5008 10000 15000 18008 20016 22008 25000 26000 27000 30000 35000 40000 45000 Number of trials to run : 4 2 2 2 2 2 2 2 2 2 1 1 1 1 1 Data alignment value (in Kbytes) : 4 4 4 4 4 4 4 4 4 4 4 1 1 1 1 Maximum memory requested that can be used=16200901024, at the size=45000 =================== Timing linear equation system solver =============== 21 Size LDA Align Time(s) GFlops Residual Residual(norm) Check 1000 1000 4 0.252 2.6533 1.071920e-12 3.655524e-02 pass 1000 1000 4 0.007 99.1793 1.071920e-12 3.655524e-02 pass 1000 1000 4 0.006 103.5554 1.071920e-12 3.655524e-02 pass 1000 1000 4 0.007 100.8084 1.071920e-12 3.655524e-02 pass 2000 2000 4 0.021 249.4091 3.866019e-12 3.362963e-02 pass 2000 2000 4 0.020 262.3845 3.866019e-12 3.362963e-02 pass 5000 5008 4 0.191 436.8663 2.298450e-11 3.205004e-02 pass 5000 5008 4 0.170 489.9396 2.298450e-11 3.205004e-02 pass 10000 10000 4 0.995 670.3465 8.869294e-11 3.127403e-02 pass 10000 10000 4 0.987 675.8754 8.869294e-11 3.127403e-02 pass Программа на OpenCL #include stdio.h #include stdlib.h #include CLcl.h #define MAX_SRC_SIZE (0x100000) // max size of kernel’s code OpenCL Vars cl_platform_id cpPlatform; // OpenCL platform cl_context cxGPUContext; // OpenCL context cl_command_queue cqCommandQueue; // OpenCL command que cl_uint uiNumDevices; // OpenCL device count cl_device_id cdDevices; // OpenCL device list cl_uint targetDevice = 0; // Default Device to compute on cl_uint uiNumDevsUsed = 1; // Number of devices used in this sample cl_program cpProgram; // OpenCL program 22 cl_kernel ckKernel; // OpenCL kernel cl_event ceEvent; // OpenCL event cl_mem cmM, cmV, cmW; // OpenCL buffers for M, V, and W size_t szGlobalWorkSize; // Total # of work items in the 1D range size_t szLocalWorkSize; // # of work items in the 1D work group size_t szParmDataBytes; // Byte size of context information size_t szKernelLength; // Byte size of kernel code cl_int ciErrNum; // Error code var char cPathAndName = NULL; // var for full paths to data, src, etc. char cSourceCL = NULL; // Buffer to hold source for compilation const char cExecutableName = NULL; Vars for BFS double ** a, * y, * x; const int N = 1000; double a[N][N] = {{}}; double y[N] = {}; // double x[N] = {}; // // number of equations // array of equations array of right side value of equations array of solution of equations 26int main() { cl_platform_id platform_id = NULL; // definition of OpenCL Vars cl_device_id device_id = NULL; 23 cl_uint ret_num_devices; cl_uint ret_num_platforms; cl_command_queue command_queue; cl_mem a_mem_obj; cl_mem b_mem_obj; cl_mem c_mem_obj; // cl_mem - type of buffers memory OpenCL cl_kernel kernel; // kernel size_t NDRange; // size of NDRange size_t work_size; // size of work-group src_str = (char )malloc(MAX_SRC_SIZE); // memory for code ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms); ret = clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_GPU, 1, &device_id, &ret_num_devices); context = clCreateContext(NULL, 1, &device_id, NULL, NULL, &ret); command_queue = clCreateCommandQueue(context, device_id, 0, &ret); a_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY, arr_size sizeof(cl_int), NULL, &ret); // create buffer for array of equations y_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY, arr_size sizeof(cl_int), NULL, &ret); // create buffer for array right side value of equations x_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY, arr_size sizeof(cl_int), NULL, &ret); // create buffer for array solutions of the equation ret = clEnqueueWriteBuffer(command_queue, a_mem_obj, CL_TRUE, 0, arr_size sizeof(int), A, 0, NULL, NULL); // write array of equations ret = clEnqueueWriteBuffer(command_queue, y_mem_obj, CL_TRUE, 0, arr_size sizeof(int), A, 0, NULL, NULL); // write array of right side value of equations program = clCreateProgramWithSource(context, 1,(const char )&src_str, (const size_t )&src_size, &ret); // create program from code ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL); // Build program kernel = clCreateKernel(program, gaussAlg, &ret); //launch kernel 27ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void )&a_mem_obj); // array of equations 24 ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void )&y_mem_obj); // array right side value of equations ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void )&x_mem_obj); // array solutions of the equation NDRange = arr_size; work_size = 64; // size of NDRange ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, &NDRange, &work_size, 0, NULL, NULL); // execute kernel ret = clFlush(command_queue); // clear the command queue ret = clFinish (command_queue); // finish execution of all commands in the queue ret = clReleaseKernel (kernel); // remove the kernel ret = clReleaseProgram (program); // remove the OpenCL program ret = clReleaseMemObject (a_mem_obj); // clear OpenCL buffer muze ret = clReleaseCommandQueue (command_queue); // delete the command queue ret = clReleaseContext (context); // remove the OpenCL context free (A); // remove local buffer muze free (Y); // remove local buffer muze free (X); // remove local buffer muze return(0); } __kernel void gaussAlg(__global const int A, __global const int Y, __global int X) // entry point { auto startTime = chronosteady_clocknow(); time start for (int i = 0; i a [i] = new double [n]; for (int j = 0; j cout << "a [" << i << "] [" << j << "] ="; cin >> a [i] [j]; } } for (int i = 0; i cout << "y [" << i << "] ="; 25 cin >> y [i]; } x = gauss (a, y, n); auto endTime = chronosteady_clocknow(); time finish long duration = chronoduration_castchronomilliseconds (endTime - startTime).count(); // amount of time spent cout << "x [" << i << "] =" << x [i] << endl; } double * gauss (double ** a, double * y, int n) // function for Gauss solution { double * x, max; int k, index; const double eps = 0.00001; // precision x = new double [n]; k = 0; while (k max = abs (a [k] [k]); // search for a string with the maximum a[i][k] index = k; for (int i = k + 1; i if (abs (a [i] [k])> max) { max = abs (a [i] [k]); index = i; } } if (max { // no non-zero diagonal elements cout << "The decision cannot be obtained because of the null column"; cout << index << "matrix A" << endl; return 0; } for (int j = 0; j double temp = a [k] [j]; a [k] [j] = a [index] [j]; a [index] [j] = temp; } double temp = y [k]; 26 y [k] = y [index]; y [index] = temp; for (int i = k; i double temp = a [i] [k]; if (abs (temp) y [i] = y [i] / temp; if (i == k) continue; // don’t subtract the equation from itself for (int j = 0; j y [i] = y [i] - y [k]; } k ++; } for (k = n - 1; k> = 0; k--) // reverse substitution { x [k] = y [k]; for (int i = 0; i } return x; } 27 |