зачем нужны cuda ядра
Что такое CUDA ядра в видеокарте
Уже долгое время технология CUDA является одной из главных особенностей видеокарт GeForce. Однако не все понимают, что это за технология и как она влияет на игры.
В этой статье расскажу и дам короткое объяснение. Так же рассмотрим и другие вопросы, которые могут возникнуть у пользователей.
Обработка графики требует одновременного выполнения сложных вычислений, именно одновременного. Поэтому в видеокартах и реализовано такое огромное количество ядер CUDA. Учитывая факт оптимизации видеокарт специально для работы с графикой, их ядра намного меньше и проще, чем у более универсальных ядер центрального процессора.
Обе технологии являются собственной разработкой компаний и в них есть различия, однако для обычного пользователя большой разницы между ними нет.
Это достаточно сложный вопрос, ответ на который не стоит искать в сухих цифрах характеристик графического адаптера. Количество не даст никаких представлений о производительности.
Многие другие характеристики, например, объем видеопамяти, поколение и скорости шины видеокарты намного важнее, для пользователя, чем данные о ядрах CUDA. Так же не стоит забывать об оптимизации в самих играх.
Лучшим способом выбора графического адаптера является все таки просмотр тестов производительности, просмотр отзывов людей, которые уже пользуются конкретной видеокартой, анализ рынка в целом, чтобы понять что выбирают покупатели. И конечно подбор по системным требованиям и fps под конкретную игру, в которую вы хотите зарубиться
Надеюсь, что помог ответить на вопрос о назначении ядер CUDA и развеять все сомнения и заблуждения о данной технологии. Теперь вы знаете что они делают и насколько важны.
CUDA: Как работает GPU
Внутренняя модель nVidia GPU – ключевой момент в понимании GPGPU с использованием CUDA. В этот раз я постараюсь наиболее детально рассказать о программном устройстве GPUs. Я расскажу о ключевых моментах компилятора CUDA, интерфейсе CUDA runtime API, ну, и в заключение, приведу пример использования CUDA для несложных математических вычислений.
Вычислительная модель GPU:
При использовании GPU вы можете задействовать грид необходимого размера и сконфигурировать блоки под нужды вашей задачи.
CUDA и язык C:
Дополнительные типы переменных и их спецификаторы будут рассмотрены непосредственно в примерах работы с памятью.
CUDA host API:
Перед тем, как приступить к непосредственному использованию CUDA для вычислений, необходимо ознакомиться с так называемым CUDA host API, который является связующим звеном между CPU и GPU. CUDA host API в свою очередь можно разделить на низкоуровневое API под названием CUDA driver API, который предоставляет доступ к драйверу пользовательского режима CUDA, и высокоуровневое API – CUDA runtime API. В своих примерах я буду использовать CUDA runtime API.
Понимаем работу GPU:
Как было сказано, нить – непосредственный исполнитель вычислений. Каким же тогда образом происходит распараллеливание вычислений между нитями? Рассмотрим работу отдельно взятого блока.
Задача. Требуется вычислить сумму двух векторов размерностью N элементов.
Нам известна максимальные размеры нашего блока: 512*512*64 нитей. Так как вектор у нас одномерный, то пока ограничимся использованием x-измерения нашего блока, то есть задействуем только одну полосу нитей из блока (рис. 3). 
Рис. 3. Наша полоса нитей из используемого блока.
Заметим, что x-размерность блока 512, то есть, мы можем сложить за один раз векторы, длина которых N // Функция сложения двух векторов
__global__ void addVector( float * left, float * right, float * result)
<
//Получаем id текущей нити.
int idx = threadIdx.x;
Таким образом, распараллеливание будет выполнено автоматически при запуске ядра. В этой функции так же используется встроенная переменная threadIdx и её поле x, которая позволяет задать соответствие между расчетом элемента вектора и нитью в блоке. Делаем расчет каждого элемента вектора в отдельной нити.
Пишем код, которые отвечает за 1 и 2 пункт в программе:
#define SIZE 512
__host__ int main()
<
//Выделяем память под вектора
float * vec1 = new float [SIZE];
float * vec2 = new float [SIZE];
float * vec3 = new float [SIZE];
//Инициализируем значения векторов
for ( int i = 0; i //Указатели на память видеокарте
float * devVec1;
float * devVec2;
float * devVec3;
…
dim3 gridSize = dim3(1, 1, 1); //Размер используемого грида
dim3 blockSize = dim3(SIZE, 1, 1); //Размер используемого блока
Теперь нам остаеться скопировать результат расчета из видеопамяти в память хоста. Но у функций ядра при этом есть особенность – асинхронное исполнение, то есть, если после вызова ядра начал работать следующий участок кода, то это ещё не значит, что GPU выполнил расчеты. Для завершения работы заданной функции ядра необходимо использовать средства синхронизации, например event’ы. Поэтому, перед копированием результатов на хост выполняем синхронизацию нитей GPU через event.
Код после вызова ядра:
//Выполняем вызов функции ядра
addVector >>(devVec1, devVec2, devVec3);
//Хендл event’а
cudaEvent_t syncEvent;
cudaEventCreate(&syncEvent); //Создаем event
cudaEventRecord(syncEvent, 0); //Записываем event
cudaEventSynchronize(syncEvent); //Синхронизируем event
Рассмотрим более подробно функции из Event Managment API.
Рис. 4. Синхронизация работы основоной и GPU прграмм.
На рисунке 4 блок «Ожидание прохождения Event’а» и есть вызов функции cudaEventSynchronize.
Ну и в заключении выводим результат на экран и чистим выделенные ресурсы.
cudaFree(devVec1);
cudaFree(devVec2);
cudaFree(devVec3);
Думаю, что описывать функции высвобождения ресурсов нет необходимости. Разве что, можно напомнить, что они так же возвращают значения cudaError_t, если есть необходимость проверки их работы.
Заключение
Надеюсь, что этот материал поможет вам понять, как функционирует GPU. Я описал самые главные моменты, которые необходимо знать для работы с CUDA. Попробуйте сами написать сложение двух матриц, но не забывайте об аппаратных ограничениях видеокарты.
CUDA: с места в карьер
NB: Статья — краткое введение, покрыть все ньюансы программирования под CUDA в одной статье вряд ли возможно 🙂
О железе
CUDA работает на видеокартых начиная с 8400GS и выше. Разные видеокарты имеют разые возможности. В целом, если вы видите что в видеокарте например 128 SP(Streaming Processor) — это значит что там 8 SIMD MP (multiprocessor), каждый из которых делает одновременно 16 операций. На один MP есть 16кб shared memory, 8192 штуки 4-хбайтных регистров (В картах серии GTX2xx значения больше). Также есть 64кб констант общие для всех MP, они кешируются, при непопадании в кеш — достаточно большая задержка (400-600 тактов). Есть глобальная память видеокарты, доступ туда не кешируется, и текстуры (кешируется, кеш оптимизирован для 2D выборок). Для использования нескольких видеокарт нужно во первый отключать SLI в дровах, а во вторых — на каждую видеокарту запускать по потоку, и вызывать cudaSetDevice().
С чего начать?
Его вы можете использовать во всех своих проектах, только вместо «../../common/inc » можно указать абсолютный путь (или переменную окружения).
nvcc — это и есть великий и ужасный компилатор CUDA. На выходе он генерирует объектный файл, в котором уже включена откомпилированная программа для видеокарты.
Обратите внимение на описание интерфейса в Mandelbrot_kernel.h — тут руками приходится описывать kernel-ы которые мы собираемся вызывать из основной С++ программы (впрочем их обычно не много, так что это не страшно).
После того как вам удалось запустить пример SDK, можно рассмотреть, чем же CUDA программа отличается от обычной.
Определение функций
Определение данных
__constant__ — задает переменную в константной памяти. Следует обратить внимание, что значения для констант нужно загружать функцией cudaMemcpyToSymbol. Константы доступны из всех тредов, скорость работы сравнима с регистрами(когда в кеш попадает).
__shared__ — задает переменную в общей памяти блока тредов (т.е. и значение будет общее на всех). Тут нужно подходить с осторожностью — компилятор агрессивно оптимизирует доступ сюда(можно придушить модификатором volatile), можно получать race condition, нужно использовать __syncthreads(); чтобы данные гарантированно записались. Shared memory разделена на банки, и когда 2 потока одновременно пытаются обратиться к одному банку, возникает bank conflict и падает скорость.
Все локальные переменные которые вы определеили в ядре (__device__) — в регистрах, самая высокая скорость доступа.
Как поток узнает над чем ему работать
Основая идея CUDA в том, что для решения вашей задачи вы запускаете тысячи и тысячи потоков, поэтому не стоит пугаться того что тут будет дальше написано 🙂
Допустим, надо сделать какую-то операцию над картинкой 200×200. Картинка разбивается на куски 10×10, и на каждый пиксел такого кусочка запускаем по потоку. Выглядить это будет так:
dim3 threads(10, 10);//размер квардатика, 10*10
dim3 grid(20, 20);//сколько квадратиков нужно чтобы покрыть все изображение
your_kernel >>(image, 200,200);//Эта строка запустит 40’000 потоков (не одновременно, одновременно работать будет 200-2000 потоков примерно).
В отличии от Brook+ от AMD, где мы сразу определяем какому потоку над какими данными работать, в CUDA все не так: передаваеиые kernel-у параметры одинаковые для всех потоков, и поток должен сам получить данные для себя, чтобы сделать это, потоку нужно вычислить, в каком месте изображения он находится. В этом помогают магические переменные blockDim, blockIdx.
const int ix = blockDim.x * blockIdx.x + threadIdx.x;
const int iy = blockDim.y * blockIdx.y + threadIdx.y;
В ix и iy — координаты, с помощью которых можно получить исходные данные из массива image, и записать результат работы.
Оптимизация
Не получается?
В первую очередь следует прочитать документацию вместе с SDK (NVIDIA_CUDA_Programming_Guide, CudaReferenceManual, ptx_isa), после этого можно спросить на официальном форуме — там даже девелоперы nVidia часто отписываются, да и вообще много умных людей. По русски можно спросить у меня на форуме например, где отвечу я 🙂 Также много людей обитает на gpgpu.ru.
Надеюсь это введение поможет людям, решившим попробовать программирование для видеокарт. Если есть проблемы/вопросы — буду рад помочь. Ну а в переди нас ждет введение в Brook+ и SIMD x86
Что такое CUDA
В статье мы подробно рассмотрим CUDA — что это за технология, как работает и в каких областях её можно использовать.
Что такое CUDA
CUDA (англ. Compute Unified Device Architecture) — это технология на базе программно-аппаратной архитектуры, которая позволяет повысить производительность параллельных вычислений. Параллельные вычисления — это вычисления, при которых процесс разработки программного обеспечения делится на потоки. Потоки обрабатываются параллельно и взаимодействуют между собой в процессе обработки. Это возможно благодаря процессорам компании NVIDIA, на которых построена работа CUDA. В их основе лежит технология GPGPU.
Прежде чем разобраться, в чём особенность этой технологии, посмотрим, какие бывают типы процессоров и как они работают.
Основные типы процессоров
CPU (англ. Central Processing Unit) — это центральный процессор, который ориентирован на обработку данных в максимально короткие сроки. Работает по следующему принципу: разбивает одну задачу на несколько потоков, каждый из которых обрабатывается отдельно. Затем эти потоки вновь объединяются и выстраиваются в нужном порядке. Такой подход позволяет увеличить скорость последовательной обработки задач. Подробнее читайте в статье Что такое процессор CPU.
GPU (англ. Graphics Processing Unit) — это специальный графический процессор, который заточен на обработку 2D или 3D-графики. Он размещается на видеокарте, что позволяет автоматически освободить основной процессор от лишней нагрузки при обработке данных. GPU состоит из нескольких тысяч ядер, которые в совокупности потребляют небольшое количество энергии. CUDA ядра «выигрывают» у CPU по производительности на 1 ватт потребляемой мощности.
На базе этих основных процессоров были разработаны различные специализированные инструменты. Например, технология GPGPU.
GPGPU (англ. General-purpose computing on graphics processing units) — это технология, которая позволяет использовать графический процессор GPU в операциях, которые обычно выполняет центральный процессор CPU. Например, в математических вычислениях. С помощью GPGPU можно использовать видеокарту для выполнения неграфических вычислений. При этом графический процессор будет работать не вместо центрального, а в качестве вычислительного блока.
CUDA является улучшенной вариацией GPGPU. Она позволяет работать на специальном диалекте — это значит, что программисты могут использовать алгоритмы, предназначенные для графических процессоров при обработке неспецифических задач. С помощью CUDA можно значимо ускорить процессы обработки. Например, можно сократить время ожидания конвертации видео до 20 раз, не задействуя основной процессор.
В каких областях можно использовать CUDA:
В последнее время технология высокоскоростной обработки становится ближе к рядовому пользователю. Например, уже существуют плагины для Photoshop, которые используют мощности GPU. С другими программами и проектами можно познакомиться на официальном сайте.
История возникновения
Первые попытки использовать графические ускорители для нецелевых вычислений предпринимались еще с конца 90-х годов. Однако масштабная работа в этом направлении началась немного позже. В начале 2000-х годов два основных лидера в мире процессоров — компании AMD и Intel – решили побороться за производство самого мощного «мозга» компьютера. В период жесткой конкуренции производителям удалось значимо поднять планку обработки процессорами тактовых частот (например, в промежуток между 2001-м и 2003-м годом частоты подросли с 1,5 до 3 ГГц).
Вместе с этим разработчики из Стенфордского университета поставили перед собой конкретную цель — создать графический адаптер, который позволит проводить неграфические вычисления. Так появился Brook — специализированный язык программирования и реализации. Затем были запущены аналогичные проекты: библиотека Accelerator, Brahma, GPU++ и другие.
Немногим позже к исследовательской работе присоединился лидер индустрии NVIDIA. Компания решила разработать собственные программные платформы и видеокарты для неграфических вычислений. Так появилась CUDA.
NVIDIA СUDA toolkit
Архитектура
Архитектура CUDA базируется на взаимодействии между CPU и GPU процессорами. Это возможно благодаря работе шейдера — программы для исполнения видеокарты процессорами.
Если говорить максимально упрощенно, то работа алгоритма CUDA выглядит следующим образом:
Взаимодействие CPU и GPU
Характеристики и среда разработки CUDA:
Недостатки и преимущества CUDA
CUDA имеет преимущества не только перед вычислениями на CPU, но и перед более ранними технологиями вычисления с помощью GPGPU. Рассмотрим, в чём они заключаются.
Как любая технология, CUDA имеет нюансы применения и ограничения.
Несмотря на все достоинства этой технологии, нельзя сказать, что параллельные вычисления полностью заменят классические. Современный центральный процессор заточен под обработку целочисленных данных и данных с плавающей запятой в максимально короткие сроки. В то же время архитектура видеокарты позволяет быстро и без проблем «распараллелить» обработку данных. Поэтому CUDA — это скорее эффективный и необходимый инструментарий для конкретных областей применения.
Теперь вы знаете, что такое и для чего нужна CUDA, а также как устроена архитектура этой технологии.
Вычисления на графических процессорах
Вычисления на графических процессорах
CUDA SDK позволяет программистам реализовывать на специальном упрощённом диалекте языка программирования Си алгоритмы, выполнимые на графических процессорах NVIDIA и включать специальные функции в текст программы на Cи. CUDA даёт разработчику возможность по своему усмотрению организовывать доступ к набору инструкций графического ускорителя и управлять его памятью, организовывать на нём сложные параллельные вычисления.
История
В 2003 г. Intel и AMD участвовали в совместной гонке за самый мощный процессор. За несколько лет в результате этой гонки тактовые частоты существенно выросли, особенно после выхода Intel Pentium 4.
После прироста тактовых частот (между 2001 и 2003 гг. тактовая частота Pentium 4 удвоилась с 1,5 до 3 ГГц), а пользователям пришлось довольствоваться десятыми долями гигагерц, которые вывели на рынок производители (с 2003 до 2005 гг.тактовые частоты увеличились 3 до 3,8 ГГц).
Работа GPU относительно простая. Она заключается в принятии группы полигонов с одной стороны и генерации группы пикселей с другой. Полигоны и пиксели независимы друг от друга, поэтому их можно обрабатывать параллельно. Таким образом, в GPU можно выделить крупную часть кристалла на вычислительные блоки, которые, в отличие от CPU, будут реально использоваться.
Первые расчёты на GPU
В дальнейшем, некоторые исследователи из проекта Brook перешли в команду разработчиков NVIDIA, чтобы представить программно-аппаратную стратегию параллельных вычислений, открыв новую долю рынка. И главным преимуществом этой инициативы NVIDIA стало то, что разработчики отлично знают все возможности своих GPU до мелочей, и в использовании графического API нет необходимости, а работать с аппаратным обеспечением можно напрямую при помощи драйвера. Результатом усилий этой команды стала NVIDIA CUDA.
Области применения параллельных расчётов на GPU
При переносе вычислений на GPU, во многих задачах достигается ускорение в 5-30 раз, по сравнению с быстрыми универсальными процессорами. Самые большие цифры (порядка 100-кратного ускорения и даже более!) достигаются на коде, который не очень хорошо подходит для расчётов при помощи блоков SSE, но вполне удобен для GPU.
Это лишь некоторые примеры ускорений синтетического кода на GPU против SSE-векторизованного кода на CPU (по данным NVIDIA):
• Флуоресцентная микроскопия: 12x.
• Молекулярная динамика (non-bonded force calc): 8-16x;
• Электростатика (прямое и многоуровневое суммирование Кулона): 40-120x и 7x.
Таблица, которую NVIDIA, показывает на всех презентациях, в которой показывается скорость графических процессоров относительно центральных.
Перечень основных приложений, в которых применяются вычисления на GPU: анализ и обработка изображений и сигналов, симуляция физики, вычислительная математика, вычислительная биология, финансовые расчёты, базы данных, динамика газов и жидкостей, криптография, адаптивная лучевая терапия, астрономия, обработка звука, биоинформатика, биологические симуляции, компьютерное зрение, анализ данных (data mining), цифровое кино и телевидение, электромагнитные симуляции, геоинформационные системы, военные применения, горное планирование, молекулярная динамика, магнитно-резонансная томография (MRI), нейросети, океанографические исследования, физика частиц, симуляция свёртывания молекул белка, квантовая химия, трассировка лучей, визуализация, радары, гидродинамическое моделирование (reservoir simulation), искусственный интеллект, анализ спутниковых данных, сейсмическая разведка, хирургия, ультразвук, видеоконференции.
Преимущества и ограничения CUDA
Чтобы перенести вычисления на GPU в рамках такой модели, нужен специальный подход. Даже поэлементное сложение двух векторов потребует отрисовки фигуры на экране или во внеэкранный буфер. Фигура растеризуется, цвет каждого пикселя вычисляется по заданной программе (пиксельному шейдеру). Программа считывает входные данные из текстур для каждого пикселя, складывает их и записывает в выходной буфер. И все эти многочисленные операции нужны для того, что в обычном языке программирования записывается одним оператором!
Программно-аппаратная архитектура для вычислений на GPU компании NVIDIA отличается от предыдущих моделей GPGPU тем, что позволяет писать программы для GPU на настоящем языке Си со стандартным синтаксисом, указателями и необходимостью в минимуме расширений для доступа к вычислительным ресурсам видеочипов. CUDA не зависит от графических API, и обладает некоторыми особенностями, предназначенными специально для вычислений общего назначения.
Преимущества CUDA перед традиционным подходом к GPGPU вычислениям
CUDA обеспечивает доступ к разделяемой между потоками памяти размером в 16 Кб на мультипроцессор, которая может быть использована для организации кэша с широкой полосой пропускания, по сравнению с текстурными выборками;
• более эффективная передача данных между системной и видеопамятью;
• отсутствие необходимости в графических API с избыточностью и накладными расходами;
• линейная адресация памяти, и gather и scatter, возможность записи по произвольным адресам;
• аппаратная поддержка целочисленных и битовых операций.
Основные ограничения CUDA:
• отсутствие поддержки рекурсии для выполняемых функций;
• минимальная ширина блока в 32 потока;
• закрытая архитектура CUDA, принадлежащая NVIDIA.
Слабыми местами программирования при помощи предыдущих методов GPGPU является то, что эти методы не используют блоки исполнения вершинных шейдеров в предыдущих неунифицированных архитектурах, данные хранятся в текстурах, а выводятся во внеэкранный буфер, а многопроходные алгоритмы используют пиксельные шейдерные блоки. В ограничения GPGPU можно включить: недостаточно эффективное использование аппаратных возможностей, ограничения полосой пропускания памяти, отсутствие операции scatter (только gather), обязательное использование графического API.
Основные преимущества CUDA по сравнению с предыдущими методами GPGPU вытекают из того, что эта архитектура спроектирована для эффективного использования неграфических вычислений на GPU и использует язык программирования C, не требуя переноса алгоритмов в удобный для концепции графического конвейера вид. CUDA предлагает новый путь вычислений на GPU, не использующий графические API, предлагающий произвольный доступ к памяти (scatter или gather). Такая архитектура лишена недостатков GPGPU и использует все исполнительные блоки, а также расширяет возможности за счёт целочисленной математики и операций битового сдвига.
CUDA открывает некоторые аппаратные возможности, недоступные из графических API, такие как разделяемая память. Это память небольшого объёма (16 килобайт на мультипроцессор), к которой имеют доступ блоки потоков. Она позволяет кэшировать наиболее часто используемые данные и может обеспечить более высокую скорость, по сравнению с использованием текстурных выборок для этой задачи. Что, в свою очередь, снижает чувствительность к пропускной способности параллельных алгоритмов во многих приложениях. Например, это полезно для линейной алгебры, быстрого преобразования Фурье и фильтров обработки изображений.
Также, графические API в обязательном порядке хранят данные в текстурах, что требует предварительной упаковки больших массивов в текстуры, что усложняет алгоритм и заставляет использовать специальную адресацию. А CUDA позволяет читать данные по любому адресу. Ещё одним преимуществом CUDA является оптимизированный обмен данными между CPU и GPU. А для разработчиков, желающих получить доступ к низкому уровню (например, при написании другого языка программирования), CUDA предлагает возможность низкоуровневого программирования на ассемблере.
Недостатки CUDA
Альтернативы CUDA
Фреймворк для написания компьютерных программ, связанных с параллельными вычислениями на различных графических и центральных процессорах. В фреймворк OpenCL входят язык программирования, который базируется на стандарте C99, и интерфейс программирования приложений (API). OpenCL обеспечивает параллелизм на уровне инструкций и на уровне данных и является реализацией техники GPGPU. OpenCL является полностью открытым стандартом, его использование не облагается лицензионными отчислениями.
Цель OpenCL состоит в том, чтобы дополнить OpenGL и OpenAL, которые являются открытыми отраслевыми стандартами для трёхмерной компьютерной графики и звука, пользуясь возможностями GPU. OpenCL разрабатывается и поддерживается некоммерческим консорциумом Khronos Group, в который входят много крупных компаний, включая Apple, AMD, Intel, nVidia, Sun Microsystems, Sony Computer Entertainment и другие.
• CAL/IL(Compute Abstraction Layer/Intermediate Language)
Областями применения ATI Stream являются приложения, требовательные к вычислительному ресурсу, такие, как финансовый анализ или обработка сейсмических данных. Использование потокового процессора позволило увеличить скорость некоторых финансовых расчётов в 55 раз по сравнению с решением той же задачи силами только центрального процессора.
NVDIA CUDA в российской научной среде.
По состоянию на декабрь 2009 г., программная модель CUDA преподается в 269 университетах мира. В России обучающие курсы по CUDA читаются в Московском, Санкт-Петербургском, Казанском, Новосибирском и Пермском государственных университетах, Международном университете природы общества и человека «Дубна», Объединённом институте ядерных исследований, Московском институте электронной техники, Ивановском государственном энергетическом университете, БГТУ им. В. Г. Шухова, МГТУ им. Баумана, РХТУ им. Менделеева, Российском научном центре «Курчатовский институт», Межрегиональном суперкомпьютерном центре РАН, Таганрогском технологическом институте (ТТИ ЮФУ).