Что такое findslide.org?

FindSlide.org - это сайт презентаций, докладов, шаблонов в формате PowerPoint.


Для правообладателей

Обратная связь

Email: Нажмите что бы посмотреть 

Яндекс.Метрика

Презентация на тему Разработка параллельных программ для GPU. Обзор CUDA API

Содержание

ОСОБЕННОСТИ CUDA APIsВиды CUDA APIs и возможности CUDA-устройств
Разработка параллельных программ для GPUОбзор CUDA API ОСОБЕННОСТИ CUDA APIsВиды CUDA APIs и возможности CUDA-устройств Виды CUDA APIsCUDA Driver APIРучная инициализация контекста GPUОтсутствуют CUDA-расширения для C++Код CPU Выбор CUDA APICUDA Driver APIБольше гибкости ( + )Сложность кода ( – Совместимость CUDA APIИмеется обратная совместимость версий Вычислительные возможности GPUCapability – это версия архитектуры CUDA GPU, которая указывает на ОЦЕНКА ПРОИЗВОДИТЕЛЬНОСТИСпособы оценки эффективности приложений CUDA Время выполненияОбщее время вычислений на GPUВремя выполнения участка кода GPU Таймеры CPUТаймеры CPU позволяют замерять общее время выполнения вычислений на GPU Таймеры CUDAТаймеры CUDA позволяют замерять время выполнения участка кода GPU Скорость передачи данныхТеоретическая пропускная способностьFDDRAM * (RDDRAM/8) * sizeof(float),где FDDRAM – частота, ОПТИМИЗАЦИЯ РАБОТЫ С ПАМЯТЬЮСпособы оптимизации работы с памятью CUDA GPU Архитектура CUDA GPU Типы памяти устройстваStreaming MultiprocessorРегистровая памятьРазделяемая памятьПамять константTexture Processing ClusterПамять текстурDDRAMЛокальная памятьГлобальная память Передача данных Host/DeviceЯвляется дорогостоящей операциейВозможна асинхронная передачаcudaMemcpy()cudaMemcpyAsync() Асинхронная передача данныхКопирование данных и выполнение ядра можно осуществлять параллельно Возможная оптимизацияСинхронная передача данных в GPUАсинхронная передача данных в GPU Нулевое копирование (Zero Copy)Прямое обращение к памяти Host’аВстроенные видеокартыИспользование кэша CPU Объединенное чтение DDRAMВыравнивание исходных данных по границе словаПотоки warp’а должны осуществлять одновременное чтение DDRAM Разделяемая память и конфликтыОбщая для всех потоков блокаРаспределяется между блокамиРазбивается на банки (32-битные слова) Регистровое давлениеРегистры жестко распределяются между потоками мультипроцессораПри большом количестве потоков возникает конфликт доступа к регистрам ВЫБОР ОПТИМАЛЬНОЙ ТОПОЛОГИИМетоды оценки топологии вычислений CUDA Степень покрытияСтепень покрытия мультипроцессора – это отношение числа активных warp'ов к максимально Определение степени покрытияCUDA GPU:8192 регистра768 потоков на мультипроцессорТопология:12 регистров на ядро128 потоков ОПТИМИЗАЦИЯ КОДАОптимизация инструкций CUDA Регистровая зависимостьИнструкция использует регистр, значение которого было получено при выполнении предыдущей инструкцииregister = instruction1(); instruction2(register); Float vs DoubleАрифметические операции с float-числами осуществляются быстрей, чем с double-числамиРекомендуется использовать Деление чиселПри делении чисел на степень двойки рекомендуется использовать оператор сдвигаX / Степень числаДля известных целых значений степеней рекомендуется использовать явное умножение вместо вызова Часто используемые функцииОбратный квадратный кореньrsqrtf() / rsqrt()Прочие арифметические операцииexpf2() / exp2()	– экспонента Точность vs СкоростьАппаратные аналоги функций__sinf() / sinf()__cosf() / cosf()__expf() / expf()Совмещенные функцииsincosf() / sincos() УПРАВЛЕНИЕ ПОТОКОМ КОМАНДОбщие рекомендации по написанию кода Операторы ветвленияИнструкции управления потоком команд (if, switch, for, while, do-while) отрицательно сказываются Предикативная запись ОТЛАДКА И ПРОФИЛИРОВАНИЕОтладка и профилирование приложений CUDA Существующие утилитыLinuxCUDA-GDB http://developer.nvidia.com/cuda-gdbWindows Vista & Windows 7NVIDIA Parallel Nsighthttp://developer.nvidia.com/nvidia-parallel-nsight АППАРАТНЫЕ ОСОБЕННОСТИ GPUКраткий обзор архитектурных особенностей GPU Причины рассогласованияОсновные причины рассогласования результатов вычислений на GPU и CPUУсечение double чисел ЛитератураNVIDIA Developer Zonehttp://developer.nvidia.com/cudaNVIDIA Parallel Nsighthttp://developer.nvidia.com/cuda-gdbCUDA C Best Practices Guidehttp://developer.download.nvidia.com/compute/cuda/4_0/toolkit/docs/CUDA_C_Best_Practices_Guide.pdf ВОПРОСЫ?
Слайды презентации

Слайд 2 ОСОБЕННОСТИ CUDA APIs
Виды CUDA APIs и возможности CUDA-устройств

ОСОБЕННОСТИ CUDA APIsВиды CUDA APIs и возможности CUDA-устройств

Слайд 3 Виды CUDA APIs
CUDA Driver API
Ручная инициализация контекста GPU
Отсутствуют

Виды CUDA APIsCUDA Driver APIРучная инициализация контекста GPUОтсутствуют CUDA-расширения для C++Код

CUDA-расширения для C++
Код CPU может компилироваться без nvcc

CUDA Runtime

API
Автоматическая инициализация контекста GPU
Наличие CUDA-расширений для C++

Слайд 4 Выбор CUDA API
CUDA Driver API
Больше гибкости ( +

Выбор CUDA APICUDA Driver APIБольше гибкости ( + )Сложность кода (

)
Сложность кода ( – )

CUDA Runtime API
Меньше гибкости (

– )
Простота кода ( + )

Слайд 5 Совместимость CUDA API
Имеется обратная совместимость версий

Совместимость CUDA APIИмеется обратная совместимость версий

Слайд 6 Вычислительные возможности GPU
Capability – это версия архитектуры CUDA

Вычислительные возможности GPUCapability – это версия архитектуры CUDA GPU, которая указывает

GPU, которая указывает на его вычислительные возможности и особенности
cudaGetDeviceProperties()


Слайд 7 ОЦЕНКА ПРОИЗВОДИТЕЛЬНОСТИ
Способы оценки эффективности приложений CUDA

ОЦЕНКА ПРОИЗВОДИТЕЛЬНОСТИСпособы оценки эффективности приложений CUDA

Слайд 8 Время выполнения
Общее время вычислений на GPU
Время выполнения участка

Время выполненияОбщее время вычислений на GPUВремя выполнения участка кода GPU

кода GPU


Слайд 9 Таймеры CPU
Таймеры CPU позволяют замерять общее время выполнения

Таймеры CPUТаймеры CPU позволяют замерять общее время выполнения вычислений на GPU

вычислений на GPU


Слайд 10 Таймеры CUDA
Таймеры CUDA позволяют замерять время выполнения участка

Таймеры CUDAТаймеры CUDA позволяют замерять время выполнения участка кода GPU

кода GPU


Слайд 11 Скорость передачи данных
Теоретическая пропускная способность

FDDRAM * (RDDRAM/8) *

Скорость передачи данныхТеоретическая пропускная способностьFDDRAM * (RDDRAM/8) * sizeof(float),где FDDRAM –

sizeof(float),
где FDDRAM – частота, RDDRAM – разрядность шины

Эффективная пропускная

способность

(BR+ BW) / time,
где BR и BW – объем прочитанной/записанной информации

Реальная пропускная способность

Слайд 12 ОПТИМИЗАЦИЯ РАБОТЫ С ПАМЯТЬЮ
Способы оптимизации работы с памятью

ОПТИМИЗАЦИЯ РАБОТЫ С ПАМЯТЬЮСпособы оптимизации работы с памятью CUDA GPU

CUDA GPU


Слайд 13 Архитектура CUDA GPU


Архитектура CUDA GPU

Слайд 14 Типы памяти устройства
Streaming Multiprocessor
Регистровая память
Разделяемая память
Память констант

Texture Processing

Типы памяти устройстваStreaming MultiprocessorРегистровая памятьРазделяемая памятьПамять константTexture Processing ClusterПамять текстурDDRAMЛокальная памятьГлобальная память

Cluster
Память текстур

DDRAM
Локальная память
Глобальная память


Слайд 15 Передача данных Host/Device
Является дорогостоящей операцией
Возможна асинхронная передача
cudaMemcpy()
cudaMemcpyAsync()

Передача данных Host/DeviceЯвляется дорогостоящей операциейВозможна асинхронная передачаcudaMemcpy()cudaMemcpyAsync()

Слайд 16 Асинхронная передача данных
Копирование данных и выполнение ядра можно

Асинхронная передача данныхКопирование данных и выполнение ядра можно осуществлять параллельно

осуществлять параллельно


Слайд 17 Возможная оптимизация
Синхронная передача данных в GPU
Асинхронная передача данных

Возможная оптимизацияСинхронная передача данных в GPUАсинхронная передача данных в GPU

в GPU


Слайд 18 Нулевое копирование (Zero Copy)
Прямое обращение к памяти Host’а

Встроенные

Нулевое копирование (Zero Copy)Прямое обращение к памяти Host’аВстроенные видеокартыИспользование кэша CPU

видеокарты
Использование кэша CPU


Слайд 19 Объединенное чтение DDRAM
Выравнивание исходных данных по границе слова

Потоки

Объединенное чтение DDRAMВыравнивание исходных данных по границе словаПотоки warp’а должны осуществлять одновременное чтение DDRAM

warp’а должны осуществлять одновременное чтение DDRAM


Слайд 20 Разделяемая память и конфликты
Общая для всех потоков блока

Распределяется

Разделяемая память и конфликтыОбщая для всех потоков блокаРаспределяется между блокамиРазбивается на банки (32-битные слова)

между блоками

Разбивается на банки (32-битные слова)


Слайд 21 Регистровое давление
Регистры жестко распределяются между потоками мультипроцессора

При большом

Регистровое давлениеРегистры жестко распределяются между потоками мультипроцессораПри большом количестве потоков возникает конфликт доступа к регистрам

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


Слайд 22 ВЫБОР ОПТИМАЛЬНОЙ ТОПОЛОГИИ
Методы оценки топологии вычислений CUDA

ВЫБОР ОПТИМАЛЬНОЙ ТОПОЛОГИИМетоды оценки топологии вычислений CUDA

Слайд 23 Степень покрытия
Степень покрытия мультипроцессора – это отношение числа

Степень покрытияСтепень покрытия мультипроцессора – это отношение числа активных warp'ов к

активных warp'ов к максимально возможному числу активных warp'ов

По количеству

используемых регистров
С учетом топологии вычислений
Без учета топологии вычислений

По размеру используемой разделяемой памяти

Слайд 24 Определение степени покрытия
CUDA GPU:
8192 регистра
768 потоков на мультипроцессор
Топология:
12

Определение степени покрытияCUDA GPU:8192 регистра768 потоков на мультипроцессорТопология:12 регистров на ядро128

регистров на ядро
128 потоков в блоке

Tmax = 8192 регистров

/ 12 регистров = 682 потока
Treal = int(682 / 128) * 128 = 640 потоков
С = Тreal / Tmax = 83%

Слайд 25 ОПТИМИЗАЦИЯ КОДА
Оптимизация инструкций CUDA

ОПТИМИЗАЦИЯ КОДАОптимизация инструкций CUDA

Слайд 26 Регистровая зависимость
Инструкция использует регистр, значение которого было получено

Регистровая зависимостьИнструкция использует регистр, значение которого было получено при выполнении предыдущей инструкцииregister = instruction1(); instruction2(register);

при выполнении предыдущей инструкции
register = instruction1(); instruction2(register);


Слайд 27 Float vs Double
Арифметические операции с float-числами осуществляются быстрей,

Float vs DoubleАрифметические операции с float-числами осуществляются быстрей, чем с double-числамиРекомендуется

чем с double-числами

Рекомендуется использовать суффикс «f» при объявлении числовых

констант, например, 3.14f

Слайд 28 Деление чисел
При делении чисел на степень двойки рекомендуется

Деление чиселПри делении чисел на степень двойки рекомендуется использовать оператор сдвигаX

использовать оператор сдвига
X / N → X >> log2(N)
X

% N → X & (N-1)

Слайд 29 Степень числа
Для известных целых значений степеней рекомендуется использовать

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

явное умножение вместо вызова pow()
pow(X, 2) → X *

X
pow(X, 3) → X * X * X

Слайд 30 Часто используемые функции
Обратный квадратный корень
rsqrtf() / rsqrt()

Прочие арифметические

Часто используемые функцииОбратный квадратный кореньrsqrtf() / rsqrt()Прочие арифметические операцииexpf2() / exp2()	–

операции
expf2() / exp2() – экспонента во 2-й степени
expf10() / exp10() –

экспонента в 10-й степени
cbrtf() / cart() – экспонента в степени 1/3
rcbrtf() / rebut() – экспонента в степени -1/3

Слайд 31 Точность vs Скорость
Аппаратные аналоги функций
__sinf() / sinf()
__cosf() /

Точность vs СкоростьАппаратные аналоги функций__sinf() / sinf()__cosf() / cosf()__expf() / expf()Совмещенные функцииsincosf() / sincos()

cosf()
__expf() / expf()

Совмещенные функции
sincosf() / sincos()


Слайд 32 УПРАВЛЕНИЕ ПОТОКОМ КОМАНД
Общие рекомендации по написанию кода

УПРАВЛЕНИЕ ПОТОКОМ КОМАНДОбщие рекомендации по написанию кода

Слайд 33 Операторы ветвления
Инструкции управления потоком команд (if, switch, for,

Операторы ветвленияИнструкции управления потоком команд (if, switch, for, while, do-while) отрицательно

while, do-while) отрицательно сказываются на производительности

В идеале все потоки

warp'а должны идти по одному пути, иначе увеличивается количество выполняемых инструкций и возможно последовательное выполнение

Слайд 34 Предикативная запись

Предикативная запись

Слайд 35 ОТЛАДКА И ПРОФИЛИРОВАНИЕ
Отладка и профилирование приложений CUDA

ОТЛАДКА И ПРОФИЛИРОВАНИЕОтладка и профилирование приложений CUDA

Слайд 36 Существующие утилиты
Linux
CUDA-GDB
http://developer.nvidia.com/cuda-gdb

Windows Vista & Windows 7
NVIDIA Parallel

Существующие утилитыLinuxCUDA-GDB http://developer.nvidia.com/cuda-gdbWindows Vista & Windows 7NVIDIA Parallel Nsighthttp://developer.nvidia.com/nvidia-parallel-nsight

Nsight
http://developer.nvidia.com/nvidia-parallel-nsight


Слайд 37 АППАРАТНЫЕ ОСОБЕННОСТИ GPU
Краткий обзор архитектурных особенностей GPU

АППАРАТНЫЕ ОСОБЕННОСТИ GPUКраткий обзор архитектурных особенностей GPU

Слайд 38 Причины рассогласования
Основные причины рассогласования результатов вычислений на GPU

Причины рассогласованияОсновные причины рассогласования результатов вычислений на GPU и CPUУсечение double

и CPU

Усечение double чисел до float при отсутствии аппаратной

поддержки double

Неассоциативность арифметических операций с дробными числами

Небольшие отклонения от стандарта IEEE 754
Особенности архитектуры процессоров x86

Слайд 39 Литература
NVIDIA Developer Zone
http://developer.nvidia.com/cuda

NVIDIA Parallel Nsight
http://developer.nvidia.com/cuda-gdb

CUDA C Best Practices

ЛитератураNVIDIA Developer Zonehttp://developer.nvidia.com/cudaNVIDIA Parallel Nsighthttp://developer.nvidia.com/cuda-gdbCUDA C Best Practices Guidehttp://developer.download.nvidia.com/compute/cuda/4_0/toolkit/docs/CUDA_C_Best_Practices_Guide.pdf

Guide
http://developer.download.nvidia.com/compute/cuda/4_0/toolkit/docs/CUDA_C_Best_Practices_Guide.pdf


  • Имя файла: razrabotka-parallelnyh-programm-dlya-gpu-obzor-cuda-api.pptx
  • Количество просмотров: 107
  • Количество скачиваний: 2