На этом шаге мы рассмотрим спецификаторы доступа.
Вводимые в CUDA расширения языка С состоят из:
- спецификаторов функций, показывающих, где будет выполнятся функция и откуда она может быть вызвана;
- спецификаторов переменных, задающих тип памяти, используемый для данных переменных;
- директивы, служащей для запуска ядра, задающей как данные, так и иерархию нитей;
- встроенных переменных, содержащих информацию о текущей нити;
- Runtime, включающей в себя дополнительные типы данных.
В CUDA используются следующие спецификаторы функций (таблица 1):
| Спецификатор | Функция выполняется из | Функция может вызываться из |
|---|---|---|
| __device__ | device(GPU) | device(GPU) |
| __global__ | device(GPU) | host(CPU) |
| __host__ | host(CPU) | host(CPU) |
Спецификаторы __host__ и __device__ могут быть использованы вместе (это значит, что соответствующая функция может выполняться как на GPU, так и на CPU - соответствующий код для обеих платформ будет автоматически сгенерирован компилятором). Спецификаторы __host__ и __global__ не могут быть использованы вместе.
Спецификатор __global__ обозначает ядро, и соответствующая функция должна возвращать значение типа void.
На функции, выполняемые на GPU (__device__ и __global__), накладываются следующие ограничения:
- нельзя брать их адрес (за исключением __global__ функций);
- не поддерживается рекурсия;
- не поддерживаются static-переменные внутри функции;
- не поддерживается переменное число входных аргументов.
Для задания размещения в памяти GPU переменных используются следующие спецификаторы:
- __device__;
- __constant__;
- __shared__.
На использование спецификаторов размещения в памяти также накладывается ряд ограничений:
- спецификаторы не могут быть применены к полям структуры;
- соответствующие переменные могут использоваться только в пределах одного файла, их нельзя объявлять, как extern;
- запись в переменные типа __constant__ может осуществляться только CPU при помощи специальных функций;
- __shared__ переменные не могут инициализироваться при объявлении.
На следующем шаге мы рассмотрим добавленные переменные.
