- PVSM.RU - https://www.pvsm.ru -
Emu [1] — это высокоуровневый язык программирования видеокарт, способный встраиваться в обычный код на системном языке программирования Rust [2].
В данной статье речь пойдёт о синтаксисе Emu, его особенностях, а также будут показаны несколько наглядных примеров его использования в реальном коде.
Cargo.toml
приведённым ниже текстом. Это вызовет скачивание последних доступных версий (если нужна конкретная сборка, то вместо *
поместите нужную версию):
[dependencies]
em = "*" // Поддержка языка Emu
ocl = "*" // Обёртка над OpenCL
Синтаксис Emu довольно прост, ведь данный язык предназначается лишь для написания функций-ядер, транслируемых в OpenCL [3] при компиляции.
Язык Emu располагает девятью типами данных, которые аналогичны соответственным в Rust. Ниже приведена таблица данных типов:
Название | Описание |
---|---|
f32 |
Тридцатидвухбитное число с плавающей точкой |
i8 |
Символ или восьмибитное число |
i16 |
Значное шестнадцатибитное число |
i32 |
Значное тридцатидвухбитное число |
i64 |
Значное шестидесятичетырехбитное число |
u8 |
Беззначное восьмибитное число |
u16 |
Беззначное шестнадцатибитное число |
u32 |
Беззначное тридцатидвухбитное число |
u64 |
Беззначное шестидесятичетырёхбитное число |
bool |
Булевое значение |
[TYPE] |
Вектор, состоящий из переменных типа TYPE |
Переменные объявляются с помощью ключевого слова let
, располагающимся за идентификатором, двоеточием, типом данных, знаком равно, присваиваемым значением и точки с запятой.
let age: i32 = 54;
let growth: f32 = 179.432;
let married: bool = true;
Конвертация примитивных типов данных осуществляется посредством бинарного оператора as
, следующим за целевым типом. Замечу, что целевым типом также может быть единица измерения (смотреть следующую секцию):
let width: i16 = 324;
let converted_width: i64 = width as i64;
Язык Emu позволяет обращаться с числами как с единицами измерения, что призвано упростить научные вычисления. В данном примере переменная length
изначально определена в метрах, но потом к ней прибавляются иные единицы измерения:
let length: f32 = 3455.345; // Метры
length += 7644.30405 as cm; // Сантиметры
length += 1687.3043 as mm; // Миллиметры
Emu располагает набором предопределённых констант, которые удобно использовать на практике. Ниже приведена соответствующая таблица.
Название | Значение |
---|---|
Y |
10 в степени 24 |
Z |
10 в степени 21 |
E |
10 в степени 18 |
P |
10 в степени 15 |
T |
10 в степени 12 |
G |
10 в степени 9 |
M |
10 в степени 6 |
k |
10 в степени 3 |
h |
10 в степени 2 |
D |
10 в степени 1 |
d |
10 в степени -1 |
c |
10 в степени -2 |
m |
10 в степени -3 |
u |
10 в степени -6 |
n |
10 в степени -9 |
p |
10 в степени -12 |
f |
10 в степени -15 |
a |
10 в степени -18 |
z |
10 в степени -21 |
y |
10 в степени -24 |
Также определены и константы, соответствующие научным данным. С таблицей, состоящей из этих постоянных, вы можете ознакомиться тут [4].
Условные операторы Emu аналогичны соответствующим операторам в Rust. Ниже показан код, применяющий условные конструкции:
let number: i32 = 2634;
let satisfied: bool = false;
if (number > 0) && (number % 2 == 0) {
satisfied = true;
}
Заголовок цикла For определяется как for NUM in START..END
, где NUM
— это переменная, принимающая значения из диапазона [START; END)
через единицу.
let sum: u64 = 0;
for i in 0..215 {
sum += i;
}
Заголовок цикла While определяется как while (CONDITION)
, где CONDITION
— это условие перехода цикла к следующей итерации. Данный код аналогичен предыдущему примеру:
let sum: u64 = 0;
let idx: i32 = 0;
while (idx < 215) {
sum += idx;
idx += 1;
}
Бесконечные циклы не имеют явно заданного условия выхода и определяются ключевым словом loop
. Они, однако, могут быть продолжены или прерваны посредством операторов break
и continue
(как и остальные два типа циклов).
let collapsed: u64 = 1;
let idx: i32 = 0;
loop {
if idx % 2 == 0 { continue; }
sum *= idx;
if idx == 12 { break; }
}
Как во всех других языках программирования, оператор return
служит выходом из текущей функции. Он также может возвращать некое значение, если сигнатура функции (смотреть следующие секции) это позволяет.
let result: i32 = 23446;
return result;
=
, +=
, -=
, *=
, /=
, %=
, &=
, ^=
, <<=
, >>=
;[IDX]
;(ARGS)
;*
для разыменования, !
для инверсии булевых данных, -
для отрицания чисел;+
, -
, *
, /
, %
, &&
, ||
, &
, |
, ^
, >>
, <<
, >
, <
, >=
, <=
, ==
, !=
.Всего есть три части функций на Emu: идентификатор, параметры и тело функции, состоящее из последовательности исполняемых инструкций. Рассмотрим функцию сложения двух чисел:
add(left f32, right f32) f32 {
return left + right;
}
Как вы уже могли заметить, данная функция возвращает сумму двух переданных в неё аргументов с помощью типа данных f32
.
Каждый параметр функции соответствует определённому адресному пространству [5]. По умолчанию, все параметры соответствуют пространству __private__
.
Добавление префиксов global_
и local_
к идентификатору параметра явно указывает его адресное пространство.
Документация [6] советует использовать префикс global_
ко всем векторам и не помечать префиксом ничего другое.
Emu предоставляет небольшой набор встроенных функций (взятых из OpenCL), позволяющих вам управлять данными GPU:
get_work_dim()
— Возвращает количество измерений;get_global_size()
— Возвращает количество глобальных элементов для заданного измерения;get_global_id()
— Возвращает уникальный идентификатор элемента для заданного измерения;get_global_size()
— Возвращает количество глобальных элементов для заданного измерения;get_local_id()
— Возвращает уникальный идентификатор локального элемента внутри конкретной рабочей группы для заданного измерения;get_num_groups()
— Возвращает количество рабочих групп для заданного измерения;get_group_id()
— Возвращает уникальный идентификатор для рабочей группы.В прикладном коде чаще всего вы встретите выражение get_global_id(0)
, возвращающее текущий индекс элемента вектора, ассоциированного с вызовом вашей функции-ядра.
Рассмотрим синтаксис вызова функций Emu из обычного кода на Rust. В качестве примера будем использовать функцию, перемножающую все элементы вектора на заданное число:
use em::emu;
emu! {
multiply(global_vector [f32], scalar f32) {
global_vector[get_global_id(0)] *= scalar;
}
}
Чтобы транслировать данную функцию в код на OpenCL, вам необходимо поместить её сигнатуру в макрос build!
следующим образом:
use em::build;
// Необходимо для макроса build! {...}
extern crate ocl;
use ocl::{flags, Platform, Device, Context, Queue, Program, Buffer, Kernel};
build! { multiply [f32] f32 }
Дальнейшие действия сводятся к вызову написанных вами функций на Emu из кода на Rust. Проще быть не может:
fn main() {
let vector = vec![0.4445, 433.245, 87.539503, 2.0];
let result = multiply(vector, 2.0).unwrap();
dbg!(result);
}
Данная программа первым аргументом принимает скаляр, на который необходимо домножить следующие аргументы. Результирующий вектор будет напечатан в консоль:
use em::{build, emu};
// Необходимо для макроса build! {...}
extern crate ocl;
use ocl::{flags, Buffer, Context, Device, Kernel, Platform, Program, Queue};
emu! {
multiply(global_vector [f32], scalar f32) {
global_vector[get_global_id(0)] *= scalar;
}
}
build! { multiply [f32] f32 }
fn main() {
// Получить все аргументы командной строки:
let args = std::env::args().collect::<Vec<String>>();
if args.len() < 3 {
panic!("Использование: cargo run -- <SCALAR> <NUMBERS>...");
}
// Скаляр должен быть указан первым аргументом:
let scalar = args[1].parse::<f32>().unwrap();
// Сконвертировать вектор строк в вектор чисел:
let vector = args[2..]
.into_iter()
.map(|string| string.parse::<f32>().unwrap())
.collect();
// Умножить и напечатать результат:
let result = multiply(vector, scalar).unwrap();
dbg!(result);
}
Выполнить данный код можно командой cargo run -- 3 2.1 3.6 6.2
. Полученный вывод соответствует ожиданиям:
[src/main.rs:33] result = [
6.2999997,
10.799999,
18.599998,
]
Надеюсь, что статья вам понравилась. Быстрый ответ на возникшие вопросы вы можете получить в русскоязычном чате [7] по языку Rust (версия для новичков [8]).
Автор: Мырзамади Темирхан
Источник [9]
Сайт-источник PVSM.RU: https://www.pvsm.ru
Путь до страницы источника: https://www.pvsm.ru/gpgpu/319789
Ссылки в тексте:
[1] Emu: https://github.com/calebwin/emu
[2] Rust: https://www.rust-lang.org/
[3] OpenCL: https://en.wikipedia.org/wiki/OpenCL
[4] тут: https://github.com/calebwin/emu/blob/master/book/numbers.md#the-emu-book
[5] определённому адресному пространству: https://www.khronos.org/registry/OpenCL/sdk/1.1/docs/man/xhtml/qualifiers.html
[6] Документация: https://github.com/calebwin/emu/blob/master/book/parameters.md#the-emu-book
[7] русскоязычном чате: https://t.me/rustlang_ru
[8] версия для новичков: https://t.me/rust_beginners_ru
[9] Источник: https://habr.com/ru/post/454678/?utm_source=habrahabr&utm_medium=rss&utm_campaign=454678
Нажмите здесь для печати.