引言

mu是一种高级图形卡编程语言，可以将其嵌入到Rust系统编程语言中的常规代码中。

本文将重点介绍Emu的语法，其功能，并显示在实际代码中使用Emu的几个说明性示例。

安装方式

1. 您要查找的库需要一个外部OpenCL依赖项。 您需要安装适合您的硬件的驱动程序。
2. Cargo.toml文本Cargo.toml 。 这将下载最新的可用版本（如果需要特定的程序集，则代替*放置所需的版本）：
   
   

    [dependencies] em = "*" //   Emu ocl = "*" //   OpenCL 

句法

Emu语法非常简单，因为该语言仅用于编写编译到OpenCL中的内核函数。

资料类型

mu语言有9种数据类型，与Rust中的数据类型相似。 下表是这些数据类型的表：

变数

使用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; 

单位

mu语言允许您将数字视为度量单位，旨在简化科学计算。 在此示例中，可变length最初以米为单位定义，但随后向其添加了其他度量单位：

 let length: f32 = 3455.345; //  length += 7644.30405 as cm; //  length += 1687.3043 as mm; //  

预定义常量

mu具有一组预定义的常数，可在实践中方便使用。 下面是相应的表。

还定义了与科学数据相对应的常数。 您可以在此处找到包含这些常量的表。

条件语句

mu条件语句类似于Rust中的相应语句。 以下代码使用条件构造：

 let number: i32 = 2634; let satisfied: bool = false; if (number > 0) && (number % 2 == 0) { satisfied = true; } 

对于循环

For循环的标题定义为for NUM in START..END中的NUM ，其中NUM是一个变量，它使用[START; END) [START; END) 。

 let sum: u64 = 0; for i in 0..215 { sum += i; } 

While循环

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返回传递给它的两个参数的总和。

地址空间

函数的每个参数都对应一个特定的地址空间 。 默认情况下，所有参数都对应于__private__空间。

在参数标识符中添加前缀global_和local_显式指示其地址空间。

该文档建议对所有向量使用global_前缀，而不对其他任何前缀添加前缀。

内建功能

mu提供了一小组内置函数（取自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) ，该表达式返回与内核函数调用关联的向量元素的当前索引。

代码执行

考虑从常规Rust代码调用Emu函数的语法。 例如，我们将使用一个函数，将向量的所有元素乘以给定的数字：

 use em::emu; emu! { multiply(global_vector [f32], scalar f32) { global_vector[get_global_id(0)] *= scalar; } } 

要将此函数转换为OpenCL代码，您需要将其签名放在build!宏中build! 如下：

 use em::build; //    build! {...} extern crate ocl; use ocl::{flags, Platform, Device, Context, Queue, Program, Buffer, Kernel}; build! { multiply [f32] f32 } 

进一步的动作归结为调用您从Rust代码编写的Emu函数。 这再简单不过了：

 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, ] 

链接到OpenCL

如前所述，Emu只是OpenCL的抽象，因此它具有与ocl crate进行交互的能力。 以下代码摘自官方存储库中的示例 ：

 use em::emu; //  "ocl"        Rust: extern crate ocl; use ocl::{flags, Platform, Device, Context, Queue, Program, Buffer, Kernel}; //  Emu    (OpenCL)   //     "EMU: &'static str": emu! { //     : multiply(global_buffer [f32], coeff f32) { global_buffer[get_global_id(0)] *= coeff; } } fn multiply(global_buffer: Vec<f32>, coeff: f32) -> ocl::Result<Vec<f32>> { //        , //  , ,   : let platform = Platform::default(); let device = Device::first(platform)?; let context = Context::builder() .platform(platform) .devices(device.clone()) .build()?; let program = Program::builder() .devices(device) .src(EMU) .build(&context)?; let queue = Queue::new(&context, device, None)?; let dims = global_buffer.len(); //    : let buffer = Buffer::<f32>::builder() .queue(queue.clone()) .flags(flags::MEM_READ_WRITE) .len(dims) .copy_host_slice(&global_buffer) .build()?; //       , //    : let kernel = Kernel::builder() .program(&program) .name("multiply") .queue(queue.clone()) .global_work_size(dims) .arg(&buffer) .arg(&coeff) .build()?; //    (    //   : unsafe { kernel.cmd() .queue(&queue) .global_work_offset(kernel.default_global_work_offset()) .global_work_size([dims, 0, 0]) .local_work_size(kernel.default_local_work_size()) .enq()?; } //  ,         // "dims": let mut vector = vec![0.0f32; dims]; buffer.cmd() .queue(&queue) .offset(0) .read(&mut vector) .enq()?; Ok(vector) } fn main() { let initial_data = vec![3.7, 4.5, 9.0, 1.2, 8.9]; //   ,   Emu,  //  "initial_data": let final_data = multiply(initial_data, 3.0).unwrap(); println!("{:?}", final_data); } 

完成时间

希望您喜欢这篇文章。 您可以在Rust（ 面向初学者的版本 ）的俄语聊天中快速找到问题的答案。