Comments 22
Большое спасибо за статью.
Пока не всё осилил, но статья интересная.
P.S.
Кстати в фотошопе можно не только nVidia задействовать, новые модели Radeon'a тоже неплохо ускоряют обработку.
P.S.
Кстати в фотошопе можно не только nVidia задействовать, новые модели Radeon'a тоже неплохо ускоряют обработку.
Ам… дык было же уже?
Сам только добавить могу, так как по работе уже применял сие =)
Перенёс один метод под видео карту… получил реальное ускорение в 30(!!!) раз
НО спешу добавить, что большой скачок получается только когда я воткнул в принципы группировки потоков в блок. и использование __constant__ и __share__ модификаторов переменных.
Самая идея там если что-то можно объекдинить в блок — объединять, тогда каждый поток может скопировать например сперва в шаред массив свою ячейку, потом синхронизироваться со всеми, и потом работать только с локальной памятью.
Чтение из обычной 400 клоков, чтение из локальной от 4х… так что вот и выйгрыш. Жаль что там только 16к на локальную память на блок, но пока и этого хватает за глаза.
все параметры, что менять не надо писать в константную область памяти (её 64к на блок)!
А так тема реальная!
Сам только добавить могу, так как по работе уже применял сие =)
Перенёс один метод под видео карту… получил реальное ускорение в 30(!!!) раз
НО спешу добавить, что большой скачок получается только когда я воткнул в принципы группировки потоков в блок. и использование __constant__ и __share__ модификаторов переменных.
Самая идея там если что-то можно объекдинить в блок — объединять, тогда каждый поток может скопировать например сперва в шаред массив свою ячейку, потом синхронизироваться со всеми, и потом работать только с локальной памятью.
Чтение из обычной 400 клоков, чтение из локальной от 4х… так что вот и выйгрыш. Жаль что там только 16к на локальную память на блок, но пока и этого хватает за глаза.
все параметры, что менять не надо писать в константную область памяти (её 64к на блок)!
А так тема реальная!
А каким образом с хоста записать данные в константную память? И не подскажите, возможна ли запись двумерных массивов в эту память?
к примеру так:
__constant__ int _threshold;
…
cudaMemcpyToSymbol( _threshold, threshold, sizeof(int));
Для работы с двумерными массивами там все есть, и специальные функции для выделения даже, чтобы выравнивание по адресам было!
Но мне всегда было с линейным удобнее работать. (если уж оптимизировать, то уметь управление над процессом подсчета индекса)
__constant__ int _threshold;
…
cudaMemcpyToSymbol( _threshold, threshold, sizeof(int));
Для работы с двумерными массивами там все есть, и специальные функции для выделения даже, чтобы выравнивание по адресам было!
Но мне всегда было с линейным удобнее работать. (если уж оптимизировать, то уметь управление над процессом подсчета индекса)
Ещё раз отдельное спасибо минусующим! я вас так люблю! =)))))))
Спасибо за статью, как раз начал разбираться с CUDA :)
А продолжение будет? )
Очень интересно как реализовывать хотя бы несколько более сложные задачи, чем Вы привели в примере. В частности интересно узнать про деление на блоки (как в этом случае нить узнает над какими данными она работает ?).
Или какой-нибудь реальный пример из жизни перевода программы с CPU на GPU.
За статью спасибо. Полезно.
Очень интересно как реализовывать хотя бы несколько более сложные задачи, чем Вы привели в примере. В частности интересно узнать про деление на блоки (как в этом случае нить узнает над какими данными она работает ?).
Или какой-нибудь реальный пример из жизни перевода программы с CPU на GPU.
За статью спасибо. Полезно.
Спасибо! Очень нужный маттериал.
MaxFX, так держать.
Однако в Вашу статью вкралась идеологическая ошибка. Размер грида равный 1 никогда не позволит выжать максимум мощи GPU при такой 100% распараллеливаемой задаче как сложение векторов. Да, формально код написан верно, но скорость работы будет на порядок меньше, чем если написать правильно. А правильно будет загрузить все стрим процессоры задачей, задав меньше размер блока и больше размер грида.
Вот например размер блока 32 треда (один варп), размер грида — 16 блоков. В таком случае разные части массива будут обрабатываться параллельно. Пока весь процесс обработки заключен в одном блоке — работает один из десятков процессоров, а остальные простаивают.
Однако в Вашу статью вкралась идеологическая ошибка. Размер грида равный 1 никогда не позволит выжать максимум мощи GPU при такой 100% распараллеливаемой задаче как сложение векторов. Да, формально код написан верно, но скорость работы будет на порядок меньше, чем если написать правильно. А правильно будет загрузить все стрим процессоры задачей, задав меньше размер блока и больше размер грида.
Вот например размер блока 32 треда (один варп), размер грида — 16 блоков. В таком случае разные части массива будут обрабатываться параллельно. Пока весь процесс обработки заключен в одном блоке — работает один из десятков процессоров, а остальные простаивают.
Картинки не грузятся
UFO just landed and posted this here
По вашему мануалу переписал свою старую программку для рисования фигур лиссажу на OpenGL… все шикарно, но… оно работает слишком быстро, рисует с такой дикой скоростью, что по экрану ползают вверх-вниз горизонтальные полосы :) Как бы это дело замедлить? :)
Картинки почините, плиз.
Sign up to leave a comment.
CUDA: Как работает GPU