eskiuCompute, graphics and machine learning on the GPU
Kernels, PBR models, LLM inference and training over WebGPU. One API for Metal, Vulkan and DirectX 12, on desktops and phones, with shaders in WGSL.
A kernel is a string of WGSL
Storage buffers, kernels and dispatches, with the results read back into Eskiu memory. Every call that can fail returns a Result with the message from WebGPU's validation layer, and a WGSL error comes with its line and column.
Underneath, <gpu/raw> is all of webgpu.h and wgpu.h, generated, with every struct's layout checked against the C compiler.
import <io>;
import <gpu>;
string twice = "@group(0) @binding(0) var<storage, read_write> v: array<f32>;\n"
"@compute @workgroup_size(64)\n"
"fn main(@builtin(global_invocation_id) id: vec3<u32>) {\n"
" v[id.x] = v[id.x] * 2.0;\n"
"}\n";
Result<int, string> run() {
float[256] data;
for (i in 0..256) data[i] = (float)i;
Gpu gpu = gpu_open()?;
defer gpu.release();
GpuBuffer v = gpu.storage(&data, sizeof(data))?;
defer v.release();
GpuKernel k = gpu.kernel(twice, "main")?;
defer k.release();
k.bind(0, &v);
k.dispatch(4, 1, 1)?;
v.read(&data, sizeof(data))?;
printf("%.1f\n", (double)data[255]); // 510.0
return Ok<int, string>(0);
}3D models, the format of AR and VR
<gpu/gltf> loads glTF 2.0 (.gltf and .glb) and <gpu/obj> Wavefront OBJ with its MTL materials, and <gpu/model> draws either with physically based shading, normal maps, alpha modes, skinning and animation.
PNG and JPEG (baseline and progressive) decode in pure Eskiu, and textures can be sRGB and mipmapped. <gpu/window> opens a window over SDL3 with its keyboard and mouse.
Gltf data = model_load("helmet.glb")?; // or a .gltf, or an .obj
defer data.free();
GpuModel model = gpu.model(&data, surface.format)?;
defer model.release();
// every frame
data.animate(0, seconds);
model.update(mat4_identity());
GpuCamera cam = model.camera(aspect);
GpuPass pass = gpu.pass(frame.view, GpuColor{0.1, 0.1, 0.1, 1.0}, &depth);
model.draw(&pass, &cam);
pass.end()?;
Machine learning on the GPU
<gpu/nn> runs transformer inference from GGUF, safetensors and llama2.c files. The Llama 2 example records the forward pass once and replays it as one compute pass per token, in f32, f16, Q8_0 and Q4_0.
| Model | Eskiu | Reference |
|---|---|---|
| stories110M (llama2.c, f16) | 350 tokens/s | llama2.c on the CPU: 22 |
| TinyLlama 1.1B Q4_0 | 178 tokens/s | llama.cpp (Metal): 260 |
| Llama 2 7B Q4_0 | 45 tokens/s | llama.cpp (Metal): 58 |
Apple M4 Pro, greedy decoding. The outputs match llama2.c byte for byte, and llama.cpp and Hugging Face transformers token for token, on the prompts tested.
Training with automatic differentiation
<gpu/train> builds a graph of operations and records the gradients of each one. The MNIST example reaches 97% test accuracy in three epochs of about 0.4 seconds, and saves the weights as safetensors that PyTorch loads with the same accuracy.
NnGraph g = gpu.nn_graph(100)?;
NnVar x = g.input(784);
NnVar labels = g.labels();
NnVar h = g.relu(g.bias(g.matmul(x, g.param(784, 128, 0.05)), g.param(1, 128, 0.0)));
NnVar logits = g.bias(g.matmul(h, g.param(128, 10, 0.125)), g.param(1, 10, 0.0));
NnVar loss = g.cross_entropy(logits, labels);
g.backward(loss)?; // the gradients of every operation
g.adam(0.001); // and the optimizer stepWhere it runs
The GPU modules go through wgpu-native, on desktops and phones. In an eskiu-lang/ui app, <gpu/ui> mounts a GpuView to draw into (a CAMetalLayer on iOS, a SurfaceView on Android). In a browser WebGPU is the browser's own, which cannot block on the GPU, so the GPU modules are planned there. The rest is plain Eskiu and runs on every target.
| Platform | Backend |
|---|---|
| macOS arm64 | Metal |
| Linux x86-64 | Vulkan |
| Windows x86-64 | DirectX 12, Vulkan |
| iOS (arm64, simulator) | Metal |
| Android (arm64, x86-64) | Vulkan |
| WebAssembly | The pure Eskiu modules (glTF, OBJ, images, GGUF, safetensors, math) today; the GPU modules are planned |
Get started
The package names wgpu-native's release archives in its manifest: the first build downloads the one for your machine, checks its SHA-256 and links it. A desktop window also needs SDL3 (brew install sdl3 on macOS). It needs Eskiu 0.14 or newer.
eskiuc new myapp && cd myapp
eskiuc add gpu --git https://github.com/eskiu-lang/gpu --rev v0.3.0
eskiuc build