rcore-os Blog
0%

GPU GPGPU GPGPGPU！

Posted on Edited on In , , , , , , , , , , , , ,
解析`WebGPU`、`Virtio1.2`规范，以及`WGPU`、`Vortex`源码，探索异步操作系统与内核态GPU资源管理的结合方案。

第四阶段总结报告

about-me: heke1228@gitee, heke1228@atom, Lfan-ke@github, heke1228@codeberg

本阶段从理解与熟悉Rust异步编程开始，探究内核态的GPU/GPGPU资源管理与异步操作系统的结合方案。

Rust异步编程

异步协程/纤程/微线程/绿色线程/虚拟线程/Future/Fiber/Promise/Coroutine/Goroutine/GreenTask/GreenThread/Microthread……名字各异（下文统一称：协程），但是表述的都是轻量级的用户态线程，挂起和恢复不涉及系统调用，开销小且灵活。使用方式在不同语言环境中大同小异，但是在实现上多多少少有不同。

Python的协程使用：

2006年通过PEP 342引入，利用生成器yield实现协程，Py3.4正式引入asyncio库，Py3.5正式协程标准化。到目前[2025.06.20 Py3.13]为止，Py协程仍然在不断发展，比如Py3.7引入的asyncio.run/create_task、Py3.11引入的async with asyncio.TaskGroup() as tg方法等等。

Py的协程源于yield生成器，目前也是可以将async def视为返回类生成器的coroutine对象。await相当于yield from。同一个线程同一时间只会运行一个协程任务队列，可以使用new_event_loop创建队列手动塞入不同的任务再使用set_event_loop管理当前活跃的任务队列，倘若开启多个协程任务队列则会直接报错。事件循环由asyncio库管理，用户直接使用高层API即可：

1
2
3
4
5
6
7
8
import asyncio

async def hello():
    print("Hello")
    await asyncio.sleep(1)
    print("World")

asyncio.run(hello())

C++的协程使用：

C++和Java的协程支持较晚，C++20才正式引入协程支持。通过co_await、co_yield和co_return关键字实现，Java可以使用SE19的虚拟线程，也可以使用子语言比如Kotlin的协程支持。C++的协程和Rust相似，依赖编译器生成状态机代码，属于无栈协程。

1
2
3
4
5
6
7
8
9
10
11
12
13
#include <cppcoro/task.hpp>
#include <cppcoro/sync_wait.hpp>
#include <iostream>

cppcoro::task<> hello() {
    std::cout << "Hello";
    co_await cppcoro::sleep_for(std::chrono::seconds(1));
    std::cout << "World";
}

int main() {
    cppcoro::sync_wait(hello());
}

JavaScript的协程使用：

之前的Js异步大多使用定时器/下帧调用实现，早期Promise形式的Promises/A规范率先在CommonJS社区流行，后续ECMA在ES6增加了Promises/A+规范的完善支持。在ES8之后正式引入了async/await语法。Js的协程是单线程事件驱动模型，通过微任务队列调度。

1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
// 早期 Promise 链
function fetchData() {
  return fetch('api/data')
    .then(response => response.json())
    .then(data => process(data))
    .catch(error => console.error(error));
}

// async/await 语法糖
async function fetchData() {
  try {
    const response = await fetch('api/data');
    const data = await response.json();
    return process(data);
  } catch (error) {
    console.error(error);
  }
}

Rust的协程使用：

Rust的协程基于Future Trait。Rust的Future得手动轮询poll函数实现才会执行。所以需要用户开发的运行时才会驱动执行。与C++类似，为无栈协程，会被编译为状态机模型，涉及唤醒模型的时候需要Awake Trait注册唤醒器，在任务均阻塞的时候避免CPU空转，而是被挂起等待被唤醒。常用的驱动库有：tokio、async-std等等。tokio 正在成为事实上的 Rust 异步运行时标准。

1
2
3
4
5
6
7
8
9
10
11
12
use tokio::time::{sleep, Duration};

async fn hello() {
    println!("Hello");
    sleep(Duration::from_secs(1)).await;
    println!("World");
}

#[tokio::main]
async fn main() {
    hello().await;
}

Rust异步运行时简易实现

由于Rust提供了Future接口，其余的调度策略等等等均由用户自定义，这样子可操作性就非常高。上述不同语言的协程实现思路均可以作为灵感来源。抛去官方的无栈协程概念不谈，也可以自己利用进程跳板的类似机制封装一个有栈协程调度器。这里实现一个简易的无栈协程调度器（暂时[2025.06]不涉及唤醒机制，优先级也是结构体多封装一个数字，使用优先队列存任务，所以只讲解最简单原型）。

目前方案及其简陋，是一个单线程的异步运行时模型，但是在合适的地方会提示多线程调度器或者其他优化的实现方案。

首先讲解Future Trait：

1
2
3
4
5
6
pub trait Future {

	type Output;
	
	fn pool(self: Pin<&mut Self>, _cx: &mut Context<'_>) -> Poll<Self::Output>;
}

所有实现Future特质的对象必须有poll惰性轮询方法。就像你管理5个小朋友，你需要将他们的作业收起来交给老师。那么你为了尽快收齐作业，你会如何去做？当然是一遍一遍一个挨着一个问：”小朋友，你的周末作业写完了吗？“。Rust的异步类似，轮询的时候只有两个状态：写完了-Poll::Ready(Output)和没写完再等等-Poll::Pending。

1
2
3
4
5
6
7
8
// 实际上在此基础上你可以封装更为复杂的轮询类型，比如：接收数据直到没有数据为止：
type Output = Option<Homework>;

match state {
	Poll::Ready(Some(homework)) => 收取,
	Poll::Ready(None) => 完成！,
	Poll::Pending => 挂起,
}

你为了方便管理这5个小朋友，你在QQ拉了一个小群，同时有一些小朋友家里有节假日活动，得过几天才能继续完成作业，你为了避免打扰他们，给他们建了另外一个小群：

1
2
let ready_queue = vec![student; 2];
let sleep_queue = vec![student; 3];

你说：”完成作业的小朋友就可以退群！当然，有活动的小朋友在活动完成之后可以加入收集作业群一起讨论作业！“

这样子一来，便成为了：你常日里可以轮询作业群的小朋友：”写完了吗？“，写完就收集作业踢出群聊。在轮询结束就在每日晚问请假群的小朋友：”接下来可以加入作业群了吗？“

1
2
3
4
5
// 上述的情况适用于单线程的轮询，为了节省CPU资源，检查sleep_queue的时候可以gap几百毫秒
// 当老师需要你检查新的小朋友的作业的时候，你就可以将其加入作业群，然后轮询：
pub fn spawn(&mut self, future: impl Future<Output = ()> + 'static) {  
    self.ready_queue.push_back(Box::pin(student));  
}

但是很快，你发现你一直在push小朋友，你自己烦，小朋友也烦，所以有没有办法让他们准备好的时候告诉你，你再去将他们移动群聊？比如，告诉小朋友的家长：”你家孩子还没写作业，办完活动告我一声，给孩子拉到作业群“。或者直接给对方父母入群二维码，当他们一家游玩结束后自己加群，这样子就不会自己一直轮询一直问了。

1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
// 那么你现在就相当于spawn了一个额外的线程，设置了一个waker
// 当满足条件的时候将会触发waker的wake方法，也就是“把孩子拉入作业群”

let waker = parent_waker();  
let mut cx = Context::from_waker(&waker);  

while let Some(student) = ready_queue.clone().iter().pop() {
	if let Poll::Ready(请假) = student.poll(&_cx) {
		// 伪代码，协助理解
		ready_queue.remove(student);
		sleep_queue.push(student);
	}
}

// 在 poll 的时候：
impl Future for Student {
	
	type Output = ();
	
	fn poll(self: Pin<&mut Self>, _cx: &mut Context<'_>) -> Poll<Self::Output> {
		
		if 活动完成 {
			Poll::Ready(())
		} else {
			// 已经告知对方父母完事了提醒我
			if self.waker_saved { return Poll::Pending; }
			// 还未告知就得告知一下
			let parent = _cx.waker().clone();
			thread::spawn(move || {
				// 他们自己花 some_duration 游玩
				thread::sleep(some_duration);
				// 游玩结束就触发他们父母提醒我拉群
				parent.wake();
			})
		}
	}
}

其中学校要统计学生节假日的行程，以确保孩子们安全，这个时候你就可以新建一个收集表，每天让孩子一家填写相关的事宜，当你发现有危险地区时就能及时阻止，或者孩子一家块回来了，就能让父母按照对应的方式提醒你：

1
2
3
4
5
6
static VTABLE: RawWakerVTable = RawWakerVTable::new(
	|data| { /* 克隆 data */ }, // clone
	|data| { /* 用 data 唤醒 */ }, // wake
	|data| { /* 引用唤醒 */ }, // wake_by_ref
	|data| { /* 释放 data */ }, // drop 
);

比如上面的你需要父母按照某个方式提醒你，或者他们自己到时候自己加群：

1
2
// 在 RawWakerVTable::new 的第二个函数参数位置写入：
|群的二维码| { 扫码加群 }

那么接下来，请假群节假日和父母出去玩的孩子加入作业群的策略就变成了：

1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
// 刚开始老师让你管理五个学生：
let mut ready_queue = [student; 5];
let mut sleep_queue = [student; 0];

// 你在作业群发现有些孩子出去玩，你将他们加入请假群：
while let Some(student) = ready_queue.clone().iter().pop() {
	if 请假出去玩 {
		// 伪代码，协助理解
		ready_queue.remove(student);
		sleep_queue.push(student);
	}
}

// 之后 poll 时告知其父母，回来后自己扫码加群
// 之后，你就会发现，当请假出去玩的回家时，就会自己加群了

接下来你只需要Poll作业群里的孩子们，让他们交作业即可了！当请加群和作业群都没人之后就是作业收齐了，就可以完成任务走人了！

后续如果想自己封装一个有特殊功能的Rust异步协程运行时可以参考：简易实现。其中如果需要异步的IO，可以基于tokio的子项目：mio进行组装，当然也可以自己基于硬件特性、操作系统特性封装唤醒机制，比如epoll、kqueue、iocp等等。

其中封装前对具体是路不是非常明确可以先行参考：利用std:net封装一个异步http客户端。此博客的思路，受益匪浅。

Rust异步爬虫的简单使用

就像py的aiohttp，rust也有自己的异步网络请求库：

1
2
3
use reqwest;  // reqwest = "0.12.15"

let response = reqwest::get(format!("https://www.baidu.com/s?wd={}", q)).await?;

之后利用tokio运行时运行异步任务即可。

Rust嵌入式异步框架介绍

Embassy是一款异步嵌入式开发框架。比RTOS更加轻量级，采用Rust的异步协程模型进行开发。其中包含一个异步执行器、一些硬件抽象层供不同板子的开发和一些异步硬件组件库：

其中，直接使用PAC层编程比较繁杂，使用HAL层抽象编程便比较轻便简单：

1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
#![no_std]
#![no_main]

use cortex_m_rt::entry;
use embassy_stm32::gpio::{Input, Level, Output, Pull, Speed};
use {defmt_rtt as _, panic_probe as _};

#[entry]
fn main() -> ! {
	let p = embassy_stm32::init(Default::default());
	let mut led = Output::new(p.PB14, Level::High, Speed::VeryHigh);
	let button = Input::new(p.PC13, Pull::Up);

	loop {
		if button.is_low() {
			led.set_high();
		} else {
			led.set_low();
		} 
	} 
}

其中Embassy的大卖点是异步框架：

1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
#![no_std]
#![no_main]
#![feature(type_alias_impl_trait)]

use embassy_executor::Spawner;
use embassy_stm32::exti::ExtiInput;
use embassy_stm32::gpio::{Input, Level, Output, Pull, Speed};
use {defmt_rtt as _, panic_probe as _};

#[embassy_executor::main]
async fn main(_spawner: Spawner) {
    let p = embassy_stm32::init(Default::default());
    let mut led = Output::new(p.PB14, Level::Low, Speed::VeryHigh);
    let mut button = ExtiInput::new(Input::new(p.PC13, Pull::Up), p.EXTI13);

    loop {
        button.wait_for_any_edge().await;
        if button.is_low() {
            led.set_high();
        } else {
            led.set_low();
        }
    }
}

补充内容：

WebGPU

src：3W规范，WGSL，WGPU，MDN

WebGPU，WWW在2021年发布WebGPU的新API，以解决上述跨平台问题，真正的跨平台框架。WebGPU是WebGL的继任者，语法类似 Rust，支持更复杂的着色器功能。比VLK更容易使用，使用WGSL作为着色器语言。可以跨平台多端使用，不仅局限于Web场景。提供更高效、灵活、安全的图形编程接口。

其中Rust依据WebGPU规范有封装框架WGPU，可以利用便捷的接口来使用GPU的计算和渲染能力。

WebGPU规范概览

WebGPU是一个提供GPU能力调用的规范接口。其中GPU嘛，目前火热的就是进行渲染-Render（比如：某3A大作震撼的特效渲染）和通用计算-GPGPU（比如：人工智能模型要在某卡上训练/推理）。所以GPU的能力大致就归类为：

  • Render Pass
  • Compute Pass

其中无论是渲染还是计算，都需要外界代码指导GPU如何进行计算，这些外界代码被称为：着色器代码。但是一定是用户可读的代码吗？不一定，比如：SPIR-V着色器代码中间表示，大部分都是面向用户的Cuda C/OCL C代码（C扩展代码），WSGL代码等等。在送向GPU的时候会被编译为GPU可执行的字节码，供GPU取址译码执行（详见第三小节-GPU架构）。

代码示例：

  • wsgl
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
struct Uniforms {
    mvpMatrix : mat4x4<f32>,
};

@binding(0) @group(0) var<uniform> uniforms : Uniforms;

struct Output {
    @builtin(position) Position : vec4<f32>,
    @location(0) vColor : vec4<f32>,
};

@vertex
fn vs_main(@location(0) pos: vec4<f32>, @location(1) color: vec4<f32>) -> Output {
    var output: Output;
    output.Position = uniforms.mvpMatrix * pos;
    output.vColor = color;
    return output;
}


@fragment
fn fs_main(@location(0) vColor: vec4<f32>) -> @location(0) vec4<f32> {
    return vColor;
}
  • opencl c
1
2
3
4
5
6
7
8
kernel void wildpointer(global uint * buffer) {

	size_t gidx = get_global_id(0);
	size_t gidy = get_global_id(1);
	size_t lidx = get_local_id(0);
	
	buffer[gidx + 4 * gidy] = (1 << gidx) | (0x10 << gidy);
}

可以看到，都需要buffer（例子1的uniforms、pos等等，例子2的buffer），而buffer一般是由CPU将数据传输到GPU的，最后的结果也可以利用数据传输指令传回。这里提到了一个非常主要的资源：buffer-缓冲区。

除了代码和缓存区外，渲染管线还可能需要以下的资源：

  • 纹理 - texture - 比如你CF枪上的皮肤/建模次世代阴影等等
  • 采样器 - sample - 决定纹理如何映射到面
  • 图形管道/计算管道 - pipeline - 渲染和计算
  • 组和布局 - bindgroup & layout - 决定数据在GPU是什么样子的，什么数据什么时候可读写

当你定义好对应的资源，以及缓冲区和代码后，就可以提交命令到管道，然后等待GPU执行渲染/计算了。

WGPU相关的介绍

src：官文，官仓，DW

WGPU是基于WebGPU规范封装的跨平台异步GPU能力调用的库。由于Rust可以非常方便的与C-BindGen/Web-WasmPack互通，WGPU可以被非常方便地跨各平台使用，安卓、手表、浏览器、小程序、桌面端、其他嵌入式设备等等等。

WGPU的项目库关系为：

1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
用户代码(JS/TS)   用户代码(Rust)
        │                │
        ▼                ▼
  deno_webgpu          wgpu
        │                │
        └───────►  wgpu-core
                         │
                      wgpu-hal
                         │
       ┌───────────┬─────┴───────┐
       ▼           ▼             |
  naga (着色器) ─ wgpu-types ──── ▼
                            底层图形API
                                 ▼
                       vlk, gles, mtl, dx12...

倘若WGPU直接编译在Web平台，则不会依赖wgpu-core，而是直接利用wasm调用WebGPU/WebGL接口。

其中用户层的vk, gles, mtl, dx12直接利用了现有的crate：ash, glow, metal, windows(winapi::um::d3d12)，而这些库大多是靠bindgen-c/o-c来绑定API的。之后被wgpu-hal统一抽象为WebGPU编程模型接口，不同的着色器语言被naga编译为中间表示后按照目前所选后端转化为对应的表示。

这里以vulkan为例，wgpu-hal使用vk::Fence+Semaphores来封装Future的fn poll来提供上层的async能力。而wgpu-core提供不安全的资源管理与交互。wgpu顶层则将wgpu-core安全化。

Rust的ash通过c-abi调用c-vk，而vk又是如何调用内核态的驱动以及如何驱动GPU设备进行计算的呢？

Vulkan Driver会自带一个加载器，通过读取特定目录的json来加载对应硬件的ICD驱动。之后通过调用符合Vulkan规范的ICD驱动提供的函数接口来驱动GPU进行渲染/计算。OpenCL也类似。

GPU架构

src：VirtioGPUv1.2规范，Vortex，MIAOW，POCL

GPU，一个熟悉又陌生的芯片。阶段四为了完成目标分析文档，实现统一内核态异步GPU计算能力资源管理驱动，理解GPU的架构是非常必要的。它和CPU类似，都有取指译码执行访存错处，也有流水线冒险分支预测等等优化手段，但是与CPU相比，究竟是什么样子的结构呢？

截图来源：北京智源智研院

截图来源：北京智源智研院

VirtioGPU简易介绍

暂时略，有缘补，可以先参考rCore-ch9的简易虚拟GPU设备

Vortex GPGPU介绍

一款基于RV架构的GPGPU。实现了OpenCL ICD及其测例，可以作为非常好的软硬一体的学习材料。

暂略，之后补，可以先看源码解析文档。

GPU驱动

远古时期的”GPU/UI”

远古时期，仅仅是一块简单的LED/LCD/OLED小屏幕，像素较小，使用颜色矩阵就可以精确的控制每一个像素的颜色，只需要板子接电使用应用层协议比如IIC传输指令即可。随着发展，每次都从计算某个像素的某点亮灭/颜色，过于麻烦，所以有了简易驱动，内部包含着绘制点线几何以及基本字体的代码和文件，此时UI编程便变成了发送指令：(x, y, w, h[, data])来控制显示。随着用户的画面需求逐渐升级，3D渲染的需求激增，英伟达推出了一个硬件支持3D渲染的显卡，后续微软、苹果等等公司也推出了相应的3D图形API，如上文提到的D3D，MTL（此时还是OGL）。

随着算力激增，绘制图形的任务被高度抽象为了渲染。此时推出的渲染引擎都接口高度化，用户不能精细控制每一个细节，比如OpenGL，而上述计算顶点与颜色的过程被抽象为：计算顶点，片段着色，光栅化，输出帧缓冲，显示。随着人工智能需求的算力激增，利用纹理存储数据、顶点变换模拟数学运算、使用帧缓冲作为输出结果：以逃课的方式使用GPU进行并行计算的大有人在，人工智能研究人员迫切需要一个“流计算”模型来并行计算大规模数据。2003年斯坦福提出BrookGPU，为GPGPU编程提供了抽象层，2006年英伟达闭源推出Cuda，此后利用GPU的并行计算能力的通用计算框架发展至今。

由于渲染绘制画面不是问题了，人们更多的开始关心如何显示的更加流畅美观，从此3D模型三角面越来越多，前端从画点画面逐渐变为了浏览器堆DIV组件、桌面端堆CMP等等。硬件变为了向人工智能通用计算助力的XPU，如：NPU等等，或者追求画质的光追RTC/RA等等。

通用计算与渲染引擎

暂略，之后补。

用户态驱动OCL介绍

暂略，之后补，可以先看适配RISC-V的POCL源码。