• Тип данных: Количество графических устройств



  • страница14/21
    Дата11.07.2018
    Размер1.74 Mb.

    Разработка алгоритмов решения систем гиперболических уравнений на графических процессорах


    1   ...   10   11   12   13   14   15   16   17   ...   21

    Распараллеливание решения 2-х мерной задачи упругости с помощью технологии CUDA на нескольких GPU


    Проблемы, с которыми мы столкнемся в разработке этого алгоритма, описаны в предыдущей главе.

    Начнем с проблемы обмена информации между устройствами.

    Есть несколько способов, первый из них, это обмениваться данными через хост. Конечно, это относительно долго. Система совершает 1 лишнее копирование. Но с другой стороны, это работает при любой конфигурации GPU устройств, т.е. не важно, как они соединены. Есть 2-ой способ, заключается он в прямом копировании с одного GPU устройства на другое (PeerToPeer Copy). PeerToPeerCopy – функция, появившаяся в CUDA 4.0 [8]. Она позволяет копировать данные с одной карты на другую, если они соединены через шину PCI-Express. Важно, что при этом не надо перегонять данные через host. Это позволяет значительно увеличить размер исходной сетки.

    Но это не всегда работает, потому что необходимым условием для такой работы является то, что все GPU подсоединены к одной шине PCI-Express. При этом мы избегаем лишнего копирования.

    Разрабатывался 2-ой вариант. Сервер содержит 8 GPU устройств (NVidia Tesla s2050). И 6 из них удовлетворяют необходимым требованиям. Поэтому все результаты, которые будут получены, исследованы на 6 GPU устройствах.

    Рассмотрим проблему синхронизации. Все вызовы функций GPU асинхронны с хостом. Однако возможность синхронизации естественно есть. Осуществляется она с помощью функции cudaDeviceSynchronize(). И должна вызываться, когда cudaSetDevice выставлена как раз на устройство, которое мы хотим синхронизировать.



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

    Рис. 5.16 Работа алгоритма на нескольких GPU.

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

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



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

    Рис. 5.17

    Каждая из матриц, изображенных на Рис. 5 .17 обрабатывается отдельным GPU, в данном случае верхняя обрабатывается GPU под номером N-1, средняя GPU под номером N, и нижняя GPU под номером N+1. Зеленым цветом обозначены данные, которые не пересчитываются на GPU, а передаются с соседнего GPU. На соседнем GPU, данные, которые были пересчитаны, и теперь должны передаваться обозначены желтым светом.

    Таким образом, мы минимизировали количество информации, передаваемой между GPU.

    Понятно, что при использовании одного GPU устройства никак не изменилась скорость работы при сравнении с алгоритмом, описанным в главе 5.4.

    Теперь рассмотрим, полученные результаты:



    График 3.12 Зависимость ускорения от количества устройств, по сравнению со временем работы на 1 устройстве для типа данных float.



    График 3.13 Зависимость ускорения от количества устройств, по сравнению со временем работы на 1 устройстве для типа данных double.

    На графиках График 3 .12 Зависимость ускорения от количества устройств, по сравнению со временем работы на 1 устройстве для типа данных float.и График 3 .13 Зависимость ускорения от количества устройств, по сравнению со временем работы на 1 устройстве для типа данных double.в легенде указаны характерные размеры сетки. Тестирование проводилось на квадратных сетках. Итак, как видно из графиков, числа немного отличаются. Но тенденция совпадает в обоих случаях. Она заключается в том, что при больших размерах сетки, мы меньше времени, относительно остальных операций, тратим на обмен данных между графическими устройствами и синхронизацию между ними и хостом, т.е. реальные вычисления на GPU становятся более тяжеловесными. А это означает, что ускорение будет расти.

    Мы получили ускорение до 4 раз на 6 карточках, этот результат можно считать приемлемым.

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


    Тип данных:

    Количество графических устройств:

    Количество памяти на каждом GPU устройстве, GB:

    Максимальный размер сетки:

    float

    6

    ~2

    17000x17000

    double

    12000x12000

    Таблица 5.3 Максимальный размер сетки при использовании нескольких GPU.

    Как результат, здесь можно зафиксировать, что мы получили максимальное количество узлов на сетки, при использовании float узлов. Что конечно уже можно использовать в практических целях.


    1   ...   10   11   12   13   14   15   16   17   ...   21

    Коьрта
    Контакты

        Главная страница


    Разработка алгоритмов решения систем гиперболических уравнений на графических процессорах