Compare commits
2 Commits
| Author | SHA1 | Date | |
|---|---|---|---|
|
e3e39af1a4 | ||
|
|
ff1226f268 |
49
README.md
49
README.md
@ -1,22 +1,37 @@
|
||||
# Minimal Zig WebGPU Compute & Render Library
|
||||
|
||||
This is a minimal, self-contained Zig library designed to simplify running compute shaders and rendering pipelines using WebGPU. It abstracts away much of the boilerplate required for GPU device initialization, memory management, bind groups, and pipeline execution.
|
||||
This is a minimal, self-contained Zig library designed to simplify running compute
|
||||
shaders and rendering pipelines using WebGPU. It abstracts away much of the boilerplate
|
||||
required for GPU device initialization, memory management, bind groups, and pipeline
|
||||
execution.
|
||||
|
||||
## Core Modules
|
||||
|
||||
The library exports the following primary components:
|
||||
|
||||
* **`GpuDevice`**: Initializes the WebGPU instance, adapter, device, and queue. It is configured to prioritize high performance and automatically requests the `ShaderF16` feature if the adapter supports it. It provides the base `GpuAllocator` for raw VRAM allocations.
|
||||
* **`GpuArenaAllocator`**: A memory management layer that wraps a base allocator to track and automatically destroy all allocated WebGPU buffers, textures, views, and pipelines when deinitialized.
|
||||
* **`GpuBuffer`**: Wraps native WebGPU buffers. It provides a `.load()` method for CPU-to-GPU data transfers and a `.read()` method that utilizes a staging buffer to map GPU data back to the CPU.
|
||||
* **`GpuCompute`**: Compiles WGSL source code into a compute pipeline and dispatches compute workgroups.
|
||||
* **`GpuRender` / `GpuTexture` / `GpuTextureView`**: Components used to initialize render pipelines, set up render attachments (textures), and bind render targets for offscreen drawing.
|
||||
* **`GpuDevice`**: Initializes the WebGPU instance, adapter, device, and queue.
|
||||
It is configured to prioritize high performance and automatically requests the `ShaderF16`
|
||||
feature if the adapter supports it. It provides the base `GpuAllocator` for raw
|
||||
VRAM allocations.
|
||||
* **`GpuArenaAllocator`**: A memory management layer that wraps a base allocator
|
||||
to track and automatically destroy all allocated WebGPU buffers, textures, views,
|
||||
and pipelines when deinitialized.
|
||||
* **`GpuBuffer`**: Wraps native WebGPU buffers. It provides a `.load()` method
|
||||
for CPU-to-GPU data transfers and a `.read()` method that utilizes a staging
|
||||
buffer to map GPU data back to the CPU.
|
||||
* **`GpuCompute`**: Compiles WGSL source code into a compute pipeline and dispatches
|
||||
compute workgroups.
|
||||
* **`GpuRender` / `GpuTexture` / `GpuTextureView`**: Components used to initialize
|
||||
render pipelines, set up render attachments (textures), and bind render targets
|
||||
for offscreen drawing.
|
||||
|
||||
---
|
||||
|
||||
## Example 1: Compute Pipeline
|
||||
|
||||
Below is a complete example demonstrating how to initialize the GPU via the device allocator, manage VRAM using a GPU Arena, run a compute shader, and read the results back to the CPU:
|
||||
Below is a complete example demonstrating how to initialize the GPU via the device
|
||||
allocator, manage VRAM using a GPU Arena, run a compute shader, and read the results
|
||||
back to the CPU:
|
||||
|
||||
```zig
|
||||
const std = @import("std");
|
||||
@ -93,7 +108,8 @@ pub fn main(init: std.process.Init) !void {
|
||||
|
||||
## Example 2: Rendering Pipeline (Offscreen to PPM Image)
|
||||
|
||||
This example demonstrates how to initialize a rendering pipeline, allocate an output texture target, draw primitives via WebGPU,
|
||||
This example demonstrates how to initialize a rendering pipeline, allocate an
|
||||
output texture target, draw primitives via WebGPU,
|
||||
and pull the frame pixels back to the CPU to write a standard image file:
|
||||
|
||||
```zig
|
||||
@ -179,7 +195,8 @@ fn savePpm(io: std.Io, filename: []const u8, w: u32, h: u32, rgba_pixels: []cons
|
||||
|
||||
## Running Examples Locally
|
||||
|
||||
If you have cloned the repository, you can run the included examples directly using the Zig build system:
|
||||
If you have cloned the repository, you can run the included examples directly
|
||||
using the Zig build system:
|
||||
|
||||
```bash
|
||||
# Run the rendering example (generates circle.ppm)
|
||||
@ -196,12 +213,14 @@ zig build bench_cp
|
||||
|
||||
## Dependencies
|
||||
|
||||
* **`wgpu.h`**: The library relies on WebGPU C API headers to bind to the native system graphics.
|
||||
* **`wgpu.h`**: The library relies on WebGPU C API headers to bind to the native
|
||||
system graphics.
|
||||
|
||||
## System Requirements
|
||||
|
||||
Because this library binds to native system graphics APIs via `wgpu-native`,
|
||||
you must ensure the appropriate development headers and libraries are available on your system before compiling.
|
||||
you must ensure the appropriate development headers and libraries are available
|
||||
on your system before compiling.
|
||||
|
||||
It work both for x86_64 and aarch64 on all platforms.
|
||||
|
||||
@ -213,11 +232,13 @@ It work both for x86_64 and aarch64 on all platforms.
|
||||
|
||||
### macOS (Metal)
|
||||
|
||||
No extra installation required. Automatically links against `Metal`, `QuartzCore`, `Foundation`, and `CoreGraphics`.
|
||||
No extra installation required. Automatically links against `Metal`, `QuartzCore`,
|
||||
`Foundation`, and `CoreGraphics`.
|
||||
|
||||
### Windows (DirectX 12)
|
||||
|
||||
No extra installation required. Automatically links against `d3d12`, `dxgi`, and `user32`. Ensure you have MSVC build tools installed.
|
||||
No extra installation required. Automatically links against `d3d12`, `dxgi`, and
|
||||
`user32`. Ensure you have MSVC build tools installed.
|
||||
|
||||
---
|
||||
|
||||
@ -226,7 +247,7 @@ No extra installation required. Automatically links against `d3d12`, `dxgi`, and
|
||||
Add it to your `build.zig.zon`:
|
||||
|
||||
```bash
|
||||
zig fetch --save git+[https://git.bouvais.lu/adrien/zig-wgpu#ref=0.2.0](https://git.bouvais.lu/adrien/zig-wgpu)
|
||||
zig fetch --save git+https://git.bouvais.lu/adrien/zig-wgpu#ref=0.3.0
|
||||
```
|
||||
|
||||
Then, expose it in your `build.zig`:
|
||||
|
||||
@ -1,6 +1,6 @@
|
||||
.{
|
||||
.name = .zig_wgpu,
|
||||
.version = "0.2.0",
|
||||
.version = "0.3.0",
|
||||
.fingerprint = 0x5d0e853acbc0c2c6,
|
||||
.minimum_zig_version = "0.16.0",
|
||||
.dependencies = .{},
|
||||
|
||||
@ -13,10 +13,10 @@ const Vec = struct {
|
||||
buf: GpuBuffer,
|
||||
len: usize,
|
||||
|
||||
// Changed: gloc is passed by value (const)
|
||||
pub fn initZero(gloc: GpuAllocator, len: usize) !Vec {
|
||||
// Changed: glloc is passed by value (const)
|
||||
pub fn initZero(glloc: GpuAllocator, len: usize) !Vec {
|
||||
return .{
|
||||
.buf = try GpuBuffer.init(gloc, .{
|
||||
.buf = try GpuBuffer.init(glloc, .{
|
||||
.size = len * @sizeOf(f16),
|
||||
.usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }),
|
||||
}),
|
||||
@ -24,9 +24,9 @@ const Vec = struct {
|
||||
};
|
||||
}
|
||||
|
||||
// Changed: gloc is passed by value
|
||||
pub fn initLoad(gloc: GpuAllocator, data: []const f16) !Vec {
|
||||
var self = try initZero(gloc, data.len);
|
||||
// Changed: glloc is passed by value
|
||||
pub fn initLoad(glloc: GpuAllocator, data: []const f16) !Vec {
|
||||
var self = try initZero(glloc, data.len);
|
||||
try self.load(data); // Direct access via the interface copy
|
||||
return self;
|
||||
}
|
||||
@ -40,18 +40,18 @@ const Vec = struct {
|
||||
try self.buf.load(f16, data);
|
||||
}
|
||||
|
||||
// Changed: gloc is passed by value instead of *GpuAllocator
|
||||
pub fn run(self: Vec, gloc: GpuAllocator, other: Vec, process: GpuCompute) !Vec {
|
||||
// Changed: glloc is passed by value instead of *GpuAllocator
|
||||
pub fn run(self: Vec, glloc: GpuAllocator, other: Vec, process: GpuCompute) !Vec {
|
||||
std.debug.assert(self.len == other.len);
|
||||
|
||||
const result = try Vec.initZero(gloc, self.len);
|
||||
const result = try Vec.initZero(glloc, self.len);
|
||||
errdefer result.deinit();
|
||||
|
||||
try process.run(gloc, .{ self.buf, other.buf, result.buf });
|
||||
try process.run(glloc, .{ self.buf, other.buf, result.buf });
|
||||
return result;
|
||||
}
|
||||
|
||||
// Changed: gloc is passed by value instead of *GpuAllocator
|
||||
// Changed: glloc is passed by value instead of *GpuAllocator
|
||||
pub fn read(self: Vec, alloc: std.mem.Allocator) ![]f16 {
|
||||
return self.buf.read(alloc, f16);
|
||||
}
|
||||
@ -63,9 +63,9 @@ pub fn main(init: std.process.Init) !void {
|
||||
|
||||
var grena = GpuArenaAllocator.init(init.gpa, device.gpuAllocator());
|
||||
defer grena.deinit();
|
||||
const gloc = grena.gpuAllocator();
|
||||
const glloc = grena.gpuAllocator();
|
||||
|
||||
const add_pip = try GpuCompute.init(gloc, @embedFile("shaders/add.wgsl"), .{ .bindings = &.{
|
||||
const add_pip = try GpuCompute.init(glloc, @embedFile("shaders/add.wgsl"), .{ .bindings = &.{
|
||||
.{ .element_size = @sizeOf(f16) },
|
||||
.{ .element_size = @sizeOf(f16) },
|
||||
.{ .element_size = @sizeOf(f16) },
|
||||
@ -120,9 +120,9 @@ pub fn main(init: std.process.Init) !void {
|
||||
// --- 1. GPU ALLOCATION PHASE ---
|
||||
const alloc_start = std.Io.Clock.awake.now(init.io);
|
||||
|
||||
const a = try Vec.initLoad(gloc, data_a);
|
||||
const a = try Vec.initLoad(glloc, data_a);
|
||||
defer a.deinit();
|
||||
const b = try Vec.initLoad(gloc, data_b);
|
||||
const b = try Vec.initLoad(glloc, data_b);
|
||||
defer b.deinit();
|
||||
|
||||
const alloc_duration = alloc_start.durationTo(std.Io.Clock.awake.now(init.io));
|
||||
@ -132,7 +132,7 @@ pub fn main(init: std.process.Init) !void {
|
||||
// --- 2. COMPUTE PHASE ---
|
||||
const compute_start = std.Io.Clock.awake.now(init.io);
|
||||
|
||||
const sum = try a.run(gloc, b, add_pip);
|
||||
const sum = try a.run(glloc, b, add_pip);
|
||||
defer sum.deinit();
|
||||
|
||||
// All 3 buffers (a, b, sum) are currently resident in VRAM here.
|
||||
|
||||
@ -20,18 +20,18 @@ pub fn main(init: std.process.Init) !void {
|
||||
// 2. Init VRAM Arena
|
||||
var grena = GpuArenaAllocator.init(allocator, device.gpuAllocator());
|
||||
defer grena.deinit();
|
||||
const gloc = grena.gpuAllocator();
|
||||
const glloc = grena.gpuAllocator();
|
||||
|
||||
// 3. Load Render Pipeline
|
||||
const circle_rp = try GpuRender.init(
|
||||
gloc,
|
||||
glloc,
|
||||
@embedFile("shaders/circle.wgsl"),
|
||||
.{ .bindings = &.{}, .texture_format = .RGBA8Unorm, .topology = .TriangleStrip },
|
||||
);
|
||||
defer circle_rp.deinit();
|
||||
|
||||
// 4. Create VRAM texture to render into
|
||||
const texture = try GpuTexture.init(gloc, .{
|
||||
const texture = try GpuTexture.init(glloc, .{
|
||||
.format = .RGBA8Unorm,
|
||||
.size = .{ .width = width, .height = height, .depthOrArrayLayers = 1 },
|
||||
.usage = .initMany(&.{ .RenderAttachment, .CopySrc }),
|
||||
@ -39,14 +39,14 @@ pub fn main(init: std.process.Init) !void {
|
||||
defer texture.deinit();
|
||||
|
||||
// 5. Create a view from texture
|
||||
const view = try GpuTextureView.init(gloc, texture, .{});
|
||||
const view = try GpuTextureView.init(glloc, texture, .{});
|
||||
defer view.deinit();
|
||||
|
||||
// 6. Run the rendering pipeline
|
||||
try circle_rp.draw(gloc, view, 4, .{});
|
||||
try circle_rp.draw(glloc, view, 4, .{});
|
||||
|
||||
// 7. Load Texture into GpuBuffer
|
||||
const cpu_staging_cpu = try texture.buffCopy(gloc);
|
||||
const cpu_staging_cpu = try texture.buffCopy(glloc);
|
||||
defer cpu_staging_cpu.deinit();
|
||||
|
||||
// 8. Read GpuBuffer to CPU
|
||||
|
||||
@ -15,11 +15,11 @@ pub fn main(init: std.process.Init) !void {
|
||||
// 2. Create a GPU Arena to manage VRAM
|
||||
var grena = GpuArenaAllocator.init(allocator, device.gpuAllocator());
|
||||
defer grena.deinit();
|
||||
const gloc = grena.gpuAllocator();
|
||||
const glloc = grena.gpuAllocator();
|
||||
|
||||
// 3. Load the WGSL compute pipeline
|
||||
const add_cp = try GpuCompute.init(
|
||||
gloc,
|
||||
glloc,
|
||||
@embedFile("shaders/add.wgsl"),
|
||||
.{
|
||||
.label = "add",
|
||||
@ -45,9 +45,9 @@ pub fn main(init: std.process.Init) !void {
|
||||
|
||||
// 5. Initialize raw GPU Buffers
|
||||
const byte_size = len * @sizeOf(f16);
|
||||
const buf_a = try GpuBuffer.init(gloc, .{ .label = "a", .size = byte_size, .usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }) });
|
||||
const buf_b = try GpuBuffer.init(gloc, .{ .label = "b", .size = byte_size, .usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }) });
|
||||
const buf_out = try GpuBuffer.init(gloc, .{ .label = "out", .size = byte_size, .usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }) });
|
||||
const buf_a = try GpuBuffer.init(glloc, .{ .label = "a", .size = byte_size, .usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }) });
|
||||
const buf_b = try GpuBuffer.init(glloc, .{ .label = "b", .size = byte_size, .usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }) });
|
||||
const buf_out = try GpuBuffer.init(glloc, .{ .label = "out", .size = byte_size, .usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }) });
|
||||
|
||||
// Note: Buffers are safely tied to the GpuArenaAllocator which will automatically
|
||||
// release them at the end. You can also manually call buf_x.deinit() if desired.
|
||||
@ -58,10 +58,10 @@ pub fn main(init: std.process.Init) !void {
|
||||
try buf_b.load(f16, data_b);
|
||||
|
||||
// 7. Dispatch the Compute
|
||||
try add_cp.run(gloc, .{ buf_a, buf_b, buf_out });
|
||||
try add_cp.run(glloc, .{ buf_a, buf_b, buf_out });
|
||||
|
||||
// 8. Map and copy the resulting buffer back to the CPU
|
||||
const staging = try GpuBuffer.init(gloc, .{
|
||||
const staging = try GpuBuffer.init(glloc, .{
|
||||
.size = byte_size,
|
||||
.usage = .initMany(&.{ .MapRead, .CopyDst }),
|
||||
});
|
||||
|
||||
@ -4,7 +4,7 @@ const GpuAllocator = @import("GpuAllocator.zig");
|
||||
const svOpt = @import("utils.zig").svOpt;
|
||||
|
||||
raw: c.WGPUBuffer,
|
||||
gloc: GpuAllocator,
|
||||
glloc: GpuAllocator,
|
||||
def: GpuBufferDef,
|
||||
|
||||
pub const GpuBufferUsage = enum(u64) {
|
||||
@ -34,12 +34,12 @@ pub const GpuBufferDef = struct {
|
||||
usage: std.EnumSet(GpuBufferUsage),
|
||||
};
|
||||
|
||||
pub fn init(gloc: GpuAllocator, def: GpuBufferDef) !@This() {
|
||||
pub fn init(glloc: GpuAllocator, def: GpuBufferDef) !@This() {
|
||||
|
||||
// Automatically align the buffer size forward to a multiple of 4 bytes under the hood
|
||||
const aligned_size = std.mem.alignForward(u64, def.size, 4);
|
||||
|
||||
const raw_handle = try gloc.allocBuffer(.{
|
||||
const raw_handle = try glloc.allocBuffer(.{
|
||||
.size = aligned_size,
|
||||
.usage = GpuBufferUsage.enumSetToWGPUBufferUsage(def.usage),
|
||||
.label = svOpt(def.label),
|
||||
@ -47,12 +47,12 @@ pub fn init(gloc: GpuAllocator, def: GpuBufferDef) !@This() {
|
||||
return .{
|
||||
.raw = raw_handle,
|
||||
.def = def,
|
||||
.gloc = gloc,
|
||||
.glloc = glloc,
|
||||
};
|
||||
}
|
||||
|
||||
pub fn deinit(self: @This()) void {
|
||||
self.gloc.freeBuffer(self.raw);
|
||||
self.glloc.freeBuffer(self.raw);
|
||||
}
|
||||
|
||||
pub fn getConstMappedRange(self: @This(), offset: u64, size: u64) ?*const anyopaque {
|
||||
@ -83,20 +83,20 @@ pub fn load(
|
||||
|
||||
if (bytes == self.def.size) {
|
||||
// Aligned path: direct download
|
||||
c.wgpuQueueWriteBuffer(self.gloc.device.queue, self.raw, 0, data.ptr, self.def.size);
|
||||
c.wgpuQueueWriteBuffer(self.glloc.device.queue, self.raw, 0, data.ptr, self.def.size);
|
||||
} else {
|
||||
// Unaligned path: Split the write into an aligned chunk and a padded remainder
|
||||
// to support arbitrary lengths without any allocations or large stack arrays.
|
||||
const aligned_part = (bytes / 4) * 4;
|
||||
if (aligned_part > 0) {
|
||||
c.wgpuQueueWriteBuffer(self.gloc.device.queue, self.raw, 0, data.ptr, aligned_part);
|
||||
c.wgpuQueueWriteBuffer(self.glloc.device.queue, self.raw, 0, data.ptr, aligned_part);
|
||||
}
|
||||
|
||||
var remainder_buf: [4]u8 = .{ 0, 0, 0, 0 };
|
||||
const data_bytes = std.mem.sliceAsBytes(data);
|
||||
@memcpy(remainder_buf[0 .. bytes - aligned_part], data_bytes[aligned_part..bytes]);
|
||||
|
||||
c.wgpuQueueWriteBuffer(self.gloc.device.queue, self.raw, aligned_part, &remainder_buf, 4);
|
||||
c.wgpuQueueWriteBuffer(self.glloc.device.queue, self.raw, aligned_part, &remainder_buf, 4);
|
||||
}
|
||||
}
|
||||
|
||||
@ -114,7 +114,7 @@ pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
||||
self.def.size,
|
||||
.{ .callback = onMapped, .userdata1 = &mapped },
|
||||
);
|
||||
while (!mapped) self.gloc.device.poll();
|
||||
while (!mapped) self.glloc.device.poll();
|
||||
|
||||
const ptr: [*]const T = @ptrCast(@alignCast(
|
||||
self.getConstMappedRange(0, self.def.size),
|
||||
@ -144,10 +144,10 @@ pub fn copy(src: @This(), dst: @This()) !void {
|
||||
if (@as(u64, GpuBufferUsage.enumSetToWGPUBufferUsage(src.def.usage)) & copy_src == 0) return error.SrcNotCopyable;
|
||||
if (@as(u64, GpuBufferUsage.enumSetToWGPUBufferUsage(dst.def.usage)) & copy_dst == 0) return error.DstNotWritable;
|
||||
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(src.gloc.device.device, null) orelse return error.Encoder;
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(src.glloc.device.device, null) orelse return error.Encoder;
|
||||
c.wgpuCommandEncoderCopyBufferToBuffer(enc, src.raw, 0, dst.raw, 0, src.def.size);
|
||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||
defer c.wgpuCommandEncoderRelease(enc);
|
||||
defer c.wgpuCommandBufferRelease(cmd);
|
||||
c.wgpuQueueSubmit(src.gloc.device.queue, 1, &cmd);
|
||||
c.wgpuQueueSubmit(src.glloc.device.queue, 1, &cmd);
|
||||
}
|
||||
|
||||
@ -22,40 +22,40 @@ pub const ComputeDef = struct {
|
||||
};
|
||||
|
||||
pip: c.WGPUComputePipeline,
|
||||
gloc: GpuAllocator,
|
||||
glloc: GpuAllocator,
|
||||
def: ComputeDef,
|
||||
|
||||
pub fn init(gloc: GpuAllocator, wgsl: []const u8, def: ComputeDef) !@This() {
|
||||
pub fn init(glloc: GpuAllocator, wgsl: []const u8, def: ComputeDef) !@This() {
|
||||
var wgsl_src = c.WGPUShaderSourceWGSL{
|
||||
.chain = .{ .sType = c.WGPUSType_ShaderSourceWGSL },
|
||||
.code = sv(wgsl),
|
||||
};
|
||||
const shader = c.wgpuDeviceCreateShaderModule(gloc.device.device, &.{
|
||||
const shader = c.wgpuDeviceCreateShaderModule(glloc.device.device, &.{
|
||||
.nextInChain = @ptrCast(&wgsl_src),
|
||||
}) orelse return error.Shader;
|
||||
defer c.wgpuShaderModuleRelease(shader);
|
||||
|
||||
const pip = try gloc.allocComputePipeline(.{
|
||||
const pip = try glloc.allocComputePipeline(.{
|
||||
.label = svOpt(def.label),
|
||||
.compute = .{ .module = shader, .entryPoint = sv("main") },
|
||||
});
|
||||
|
||||
return .{
|
||||
.gloc = gloc,
|
||||
.glloc = glloc,
|
||||
.pip = pip,
|
||||
.def = def,
|
||||
};
|
||||
}
|
||||
|
||||
pub fn deinit(self: @This()) void {
|
||||
self.gloc.freeComputePipeline(self.pip);
|
||||
self.glloc.freeComputePipeline(self.pip);
|
||||
}
|
||||
|
||||
/// Execute the compute pass with arbitrary buffer bindings via a tuple.
|
||||
/// Example: `try proc.run(gloc, .{ buf_a, buf_b, buf_out });`
|
||||
/// Example: `try proc.run(glloc, .{ buf_a, buf_b, buf_out });`
|
||||
pub fn run(
|
||||
self: @This(),
|
||||
gloc: GpuAllocator,
|
||||
glloc: GpuAllocator,
|
||||
args: anytype,
|
||||
) !void {
|
||||
const type_info = @typeInfo(@TypeOf(args));
|
||||
@ -113,12 +113,12 @@ pub fn run(
|
||||
defer if (info_buf) |b| b.deinit();
|
||||
|
||||
if (self.def.append_info_buffer) {
|
||||
info_buf = try GpuBuffer.init(gloc, .{
|
||||
info_buf = try GpuBuffer.init(glloc, .{
|
||||
.size = @sizeOf(u32),
|
||||
.usage = .initMany(&.{ .Uniform, .CopyDst }),
|
||||
.label = "compute_info_buffer",
|
||||
});
|
||||
c.wgpuQueueWriteBuffer(gloc.device.queue, info_buf.?.raw, 0, &elements_count, @sizeOf(u32));
|
||||
c.wgpuQueueWriteBuffer(glloc.device.queue, info_buf.?.raw, 0, &elements_count, @sizeOf(u32));
|
||||
|
||||
entries_buf[entry_count] = .{
|
||||
.binding = @intCast(entry_count),
|
||||
@ -130,11 +130,11 @@ pub fn run(
|
||||
}
|
||||
|
||||
const entries = entries_buf[0..entry_count];
|
||||
try submitPass(gloc, self.pip, entries, elements_count, self.def.workgroup_size, self.def.max_workgroups);
|
||||
try submitPass(glloc, self.pip, entries, elements_count, self.def.workgroup_size, self.def.max_workgroups);
|
||||
}
|
||||
|
||||
fn submitPass(
|
||||
gloc: GpuAllocator,
|
||||
glloc: GpuAllocator,
|
||||
pipeline: c.WGPUComputePipeline,
|
||||
entries: []const c.WGPUBindGroupEntry,
|
||||
n: usize,
|
||||
@ -146,14 +146,14 @@ fn submitPass(
|
||||
const bgl = c.wgpuComputePipelineGetBindGroupLayout(pipeline, 0);
|
||||
defer c.wgpuBindGroupLayoutRelease(bgl);
|
||||
|
||||
const bg = c.wgpuDeviceCreateBindGroup(gloc.device.device, &.{
|
||||
const bg = c.wgpuDeviceCreateBindGroup(glloc.device.device, &.{
|
||||
.layout = bgl,
|
||||
.entries = entries.ptr,
|
||||
.entryCount = entries.len,
|
||||
}) orelse return error.BindGroup;
|
||||
defer c.wgpuBindGroupRelease(bg);
|
||||
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(gloc.device.device, null) orelse return error.Encoder;
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(glloc.device.device, null) orelse return error.Encoder;
|
||||
const pass = c.wgpuCommandEncoderBeginComputePass(enc, null);
|
||||
c.wgpuComputePassEncoderSetPipeline(pass, pipeline);
|
||||
c.wgpuComputePassEncoderSetBindGroup(pass, 0, bg, 0, null);
|
||||
@ -168,7 +168,7 @@ fn submitPass(
|
||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||
defer c.wgpuCommandEncoderRelease(enc);
|
||||
defer c.wgpuCommandBufferRelease(cmd);
|
||||
c.wgpuQueueSubmit(gloc.device.queue, 1, &cmd);
|
||||
c.wgpuQueueSubmit(glloc.device.queue, 1, &cmd);
|
||||
}
|
||||
|
||||
fn ceilDiv(n: usize, d: usize) usize {
|
||||
|
||||
@ -34,16 +34,16 @@ const GpuPrimitiveTopology = enum(c_uint) {
|
||||
Force32 = 0x7FFFFFFF,
|
||||
};
|
||||
|
||||
gloc: GpuAllocator,
|
||||
glloc: GpuAllocator,
|
||||
pip: c.WGPURenderPipeline,
|
||||
def: GpuRenderDef,
|
||||
|
||||
pub fn init(gloc: GpuAllocator, wgsl: []const u8, def: GpuRenderDef) !@This() {
|
||||
pub fn init(glloc: GpuAllocator, wgsl: []const u8, def: GpuRenderDef) !@This() {
|
||||
var wgsl_src = c.WGPUShaderSourceWGSL{
|
||||
.chain = .{ .sType = c.WGPUSType_ShaderSourceWGSL },
|
||||
.code = sv(wgsl),
|
||||
};
|
||||
const shader = c.wgpuDeviceCreateShaderModule(gloc.device.device, &.{
|
||||
const shader = c.wgpuDeviceCreateShaderModule(glloc.device.device, &.{
|
||||
.nextInChain = @ptrCast(&wgsl_src),
|
||||
}) orelse return error.Shader;
|
||||
defer c.wgpuShaderModuleRelease(shader);
|
||||
@ -69,7 +69,7 @@ pub fn init(gloc: GpuAllocator, wgsl: []const u8, def: GpuRenderDef) !@This() {
|
||||
};
|
||||
|
||||
// 3. Compile the Complete Render Pipeline
|
||||
const pip = try gloc.allocRenderPipeline(.{
|
||||
const pip = try glloc.allocRenderPipeline(.{
|
||||
.label = svOpt(def.label),
|
||||
.vertex = .{
|
||||
.module = shader,
|
||||
@ -90,21 +90,21 @@ pub fn init(gloc: GpuAllocator, wgsl: []const u8, def: GpuRenderDef) !@This() {
|
||||
});
|
||||
|
||||
return .{
|
||||
.gloc = gloc,
|
||||
.glloc = glloc,
|
||||
.pip = pip,
|
||||
.def = def,
|
||||
};
|
||||
}
|
||||
|
||||
pub fn deinit(self: @This()) void {
|
||||
self.gloc.freeRenderPipeline(self.pip);
|
||||
self.glloc.freeRenderPipeline(self.pip);
|
||||
}
|
||||
|
||||
/// Execute the render pass targeting a specific frame texture view.
|
||||
/// Passes bind groups via a tuple exactly like your original compute setup.
|
||||
pub fn draw(
|
||||
self: @This(),
|
||||
gloc: GpuAllocator,
|
||||
glloc: GpuAllocator,
|
||||
target_view: GpuTextureView,
|
||||
vertex_count: u32,
|
||||
args: anytype,
|
||||
@ -138,7 +138,7 @@ pub fn draw(
|
||||
const bgl = c.wgpuRenderPipelineGetBindGroupLayout(self.pip, 0);
|
||||
defer c.wgpuBindGroupLayoutRelease(bgl);
|
||||
|
||||
const bg = c.wgpuDeviceCreateBindGroup(gloc.device.device, &.{
|
||||
const bg = c.wgpuDeviceCreateBindGroup(glloc.device.device, &.{
|
||||
.layout = bgl,
|
||||
.entries = entries.ptr,
|
||||
.entryCount = @intCast(entries.len),
|
||||
@ -146,7 +146,7 @@ pub fn draw(
|
||||
defer c.wgpuBindGroupRelease(bg);
|
||||
|
||||
// Encode Render Command
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(gloc.device.device, null) orelse return error.Encoder;
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(glloc.device.device, null) orelse return error.Encoder;
|
||||
defer c.wgpuCommandEncoderRelease(enc);
|
||||
|
||||
const color_attachment = c.WGPURenderPassColorAttachment{
|
||||
@ -180,5 +180,5 @@ pub fn draw(
|
||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||
defer c.wgpuCommandBufferRelease(cmd);
|
||||
|
||||
c.wgpuQueueSubmit(gloc.device.queue, 1, &cmd);
|
||||
c.wgpuQueueSubmit(glloc.device.queue, 1, &cmd);
|
||||
}
|
||||
|
||||
@ -14,10 +14,10 @@ pub const GpuTextureDef = struct {
|
||||
};
|
||||
|
||||
raw: c.WGPUTexture,
|
||||
gloc: GpuAllocator,
|
||||
glloc: GpuAllocator,
|
||||
def: GpuTextureDef,
|
||||
|
||||
pub fn init(gloc: GpuAllocator, def: GpuTextureDef) !@This() {
|
||||
pub fn init(glloc: GpuAllocator, def: GpuTextureDef) !@This() {
|
||||
var use: u64 = 0;
|
||||
var iter = def.usage.iterator();
|
||||
while (iter.next()) |flag| use |= @intFromEnum(flag);
|
||||
@ -31,13 +31,13 @@ pub fn init(gloc: GpuAllocator, def: GpuTextureDef) !@This() {
|
||||
.mipLevelCount = 1,
|
||||
.sampleCount = 1,
|
||||
};
|
||||
const raw = try gloc.allocTexture(desc);
|
||||
const raw = try glloc.allocTexture(desc);
|
||||
|
||||
return .{ .gloc = gloc, .raw = raw, .def = def };
|
||||
return .{ .glloc = glloc, .raw = raw, .def = def };
|
||||
}
|
||||
|
||||
pub fn deinit(self: @This()) void {
|
||||
self.gloc.freeTexture(self.raw);
|
||||
self.glloc.freeTexture(self.raw);
|
||||
}
|
||||
|
||||
pub fn getConstMappedRange(self: @This(), offset: u64, size: u64) ?*const anyopaque {
|
||||
@ -53,14 +53,14 @@ pub fn bytesSizeRow(self: @This()) u32 {
|
||||
}
|
||||
|
||||
/// Return a GpuBuffer containing a copy of the texture.
|
||||
pub fn buffCopy(self: @This(), gloc: GpuAllocator) !GpuBuffer {
|
||||
const buf = try GpuBuffer.init(gloc, .{
|
||||
pub fn buffCopy(self: @This(), glloc: GpuAllocator) !GpuBuffer {
|
||||
const buf = try GpuBuffer.init(glloc, .{
|
||||
.size = self.bytesSize(),
|
||||
.usage = .initMany(&.{ .CopyDst, .CopySrc }),
|
||||
.label = "texture_copy_buffer",
|
||||
});
|
||||
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(gloc.device.device, null) orelse return error.Encoder;
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(glloc.device.device, null) orelse return error.Encoder;
|
||||
defer c.wgpuCommandEncoderRelease(enc);
|
||||
|
||||
const src_copy = c.WGPUTexelCopyTextureInfo{
|
||||
@ -82,7 +82,7 @@ pub fn buffCopy(self: @This(), gloc: GpuAllocator) !GpuBuffer {
|
||||
|
||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||
defer c.wgpuCommandBufferRelease(cmd);
|
||||
c.wgpuQueueSubmit(gloc.device.queue, 1, &cmd);
|
||||
c.wgpuQueueSubmit(glloc.device.queue, 1, &cmd);
|
||||
|
||||
return buf;
|
||||
}
|
||||
@ -110,7 +110,7 @@ pub fn load(
|
||||
const bytes = data.len * @sizeOf(T);
|
||||
|
||||
c.wgpuQueueWriteTexture(
|
||||
self.gloc.device.queue,
|
||||
self.glloc.device.queue,
|
||||
&.{
|
||||
.texture = self.raw,
|
||||
.mipLevel = 0,
|
||||
@ -132,14 +132,14 @@ pub fn load(
|
||||
pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
||||
const out = try alloc.alloc(T, @divExact(self.size, @sizeOf(T)));
|
||||
|
||||
const staging = try init(self.gloc, .{
|
||||
const staging = try init(self.glloc, .{
|
||||
.size = self.size,
|
||||
.usage = .initMany(&.{ .MapRead, .CopyDst }),
|
||||
.label = "texture_read_staging",
|
||||
});
|
||||
defer staging.deinit();
|
||||
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(self.gloc.device.device, null) orelse return error.Encoder;
|
||||
const enc = c.wgpuDeviceCreateCommandEncoder(self.glloc.device.device, null) orelse return error.Encoder;
|
||||
const src_copy = c.WGPUTexelCopyTextureInfo{
|
||||
.texture = self.raw,
|
||||
.mipLevel = 0,
|
||||
@ -158,7 +158,7 @@ pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||
defer c.wgpuCommandEncoderRelease(enc);
|
||||
defer c.wgpuCommandBufferRelease(cmd);
|
||||
c.wgpuQueueSubmit(self.gloc.device.queue, 1, &cmd);
|
||||
c.wgpuQueueSubmit(self.glloc.device.queue, 1, &cmd);
|
||||
|
||||
var mapped = false;
|
||||
staging.mapAsync(
|
||||
@ -167,7 +167,7 @@ pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
||||
self.size,
|
||||
.{ .callback = onMapped, .userdata1 = &mapped },
|
||||
);
|
||||
while (!mapped) self.gloc.device.poll();
|
||||
while (!mapped) self.glloc.device.poll();
|
||||
|
||||
const ptr: [*]const T = @ptrCast(@alignCast(
|
||||
staging.getConstMappedRange(0, self.size),
|
||||
|
||||
@ -13,23 +13,23 @@ pub const GpuViewDef = struct {
|
||||
};
|
||||
|
||||
raw: c.WGPUTextureView,
|
||||
gloc: GpuAllocator,
|
||||
glloc: GpuAllocator,
|
||||
|
||||
pub fn init(gloc: GpuAllocator, texture: GpuTexture, def: GpuViewDef) !@This() {
|
||||
pub fn init(glloc: GpuAllocator, texture: GpuTexture, def: GpuViewDef) !@This() {
|
||||
var use: u64 = 0;
|
||||
var iter = def.usage.iterator();
|
||||
while (iter.next()) |flag| use |= @intFromEnum(flag);
|
||||
|
||||
const raw = try gloc.allocTextureView(texture.raw, .{
|
||||
const raw = try glloc.allocTextureView(texture.raw, .{
|
||||
.label = svOpt(def.label),
|
||||
.format = @intFromEnum(def.format),
|
||||
.usage = use,
|
||||
.mipLevelCount = 1,
|
||||
.arrayLayerCount = 1,
|
||||
});
|
||||
return .{ .gloc = gloc, .raw = raw };
|
||||
return .{ .glloc = glloc, .raw = raw };
|
||||
}
|
||||
|
||||
pub fn deinit(self: @This()) void {
|
||||
self.gloc.freeTextureView(self.raw);
|
||||
self.glloc.freeTextureView(self.raw);
|
||||
}
|
||||
|
||||
Loading…
x
Reference in New Issue
Block a user