Compare commits
3 Commits
28f92570e8
...
389060c238
| Author | SHA1 | Date | |
|---|---|---|---|
|
|
389060c238 | ||
|
|
a75252b68f | ||
|
|
d010b37801 |
62
README.md
62
README.md
@ -8,7 +8,7 @@ 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.
|
* **`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.
|
* **`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.
|
* **`GpuBuffer`**: Wraps native WebGPU buffers. It provides a `.load()` method for CPU-to-GPU data transfers and a `.read()` method to map GPU data back to the CPU.
|
||||||
* **`GpuCompute`**: Compiles WGSL source code into a compute pipeline and dispatches compute workgroups.
|
* **`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.
|
* **`GpuRender` / `GpuTexture` / `GpuTextureView`**: Components used to initialize render pipelines, set up render attachments (textures), and bind render targets for offscreen drawing.
|
||||||
|
|
||||||
@ -36,17 +36,20 @@ pub fn main(init: std.process.Init) !void {
|
|||||||
// 2. Create a GPU Arena to manage VRAM
|
// 2. Create a GPU Arena to manage VRAM
|
||||||
var grena = GpuArenaAllocator.init(allocator, device.gpuAllocator());
|
var grena = GpuArenaAllocator.init(allocator, device.gpuAllocator());
|
||||||
defer grena.deinit();
|
defer grena.deinit();
|
||||||
const gloc = grena.gpuAllocator();
|
const glloc = grena.gpuAllocator();
|
||||||
|
|
||||||
// 3. Load the WGSL compute pipeline
|
// 3. Load the WGSL compute pipeline
|
||||||
const add_cp = try GpuCompute.init(
|
const add_cp = try GpuCompute.init(
|
||||||
gloc,
|
glloc,
|
||||||
@embedFile("shaders/add.wgsl"),
|
@embedFile("shaders/add.wgsl"),
|
||||||
.{ .bindings = &.{
|
.{
|
||||||
.{ .element_size = @sizeOf(f16) },
|
.label = "add",
|
||||||
.{ .element_size = @sizeOf(f16) },
|
.bindings = &.{
|
||||||
.{ .element_size = @sizeOf(f16) },
|
.{ .element_size = @sizeOf(f16) },
|
||||||
} },
|
.{ .element_size = @sizeOf(f16) },
|
||||||
|
.{ .element_size = @sizeOf(f16) },
|
||||||
|
},
|
||||||
|
},
|
||||||
);
|
);
|
||||||
|
|
||||||
// 4. Setup CPU data
|
// 4. Setup CPU data
|
||||||
@ -63,19 +66,20 @@ pub fn main(init: std.process.Init) !void {
|
|||||||
|
|
||||||
// 5. Initialize raw GPU Buffers
|
// 5. Initialize raw GPU Buffers
|
||||||
const byte_size = len * @sizeOf(f16);
|
const byte_size = len * @sizeOf(f16);
|
||||||
const buf_a = try GpuBuffer.init(gloc, byte_size, .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(gloc, byte_size, .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(gloc, byte_size, .initMany(&.{ .Storage, .CopyDst, .CopySrc }));
|
const buf_out = try GpuBuffer.init(glloc, .{ .label = "out", .size = byte_size, .usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }) });
|
||||||
|
|
||||||
// Note: Buffers, pipelines, and other objects initialized with 'gloc'
|
// Note: Buffers are safely tied to the GpuArenaAllocator which will automatically
|
||||||
// are safely tied to the GpuArenaAllocator and will automatically release.
|
// release them at the end. You can also manually call buf_x.deinit() if desired.
|
||||||
|
// This will also release pipelines, textures, ect. Everything using a GpuAllocator to init.
|
||||||
|
|
||||||
// 6. Transfer data from CPU slices to GPU Buffers
|
// 6. Transfer data from CPU slices to GPU Buffers
|
||||||
try buf_a.load(f16, data_a);
|
try buf_a.load(f16, data_a);
|
||||||
try buf_b.load(f16, data_b);
|
try buf_b.load(f16, data_b);
|
||||||
|
|
||||||
// 7. Dispatch the Compute
|
// 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
|
// 8. Map and copy the resulting buffer back to the CPU
|
||||||
const out = try buf_out.read(allocator, f16);
|
const out = try buf_out.read(allocator, f16);
|
||||||
@ -96,6 +100,7 @@ and pull the frame pixels back to the CPU to write a standard image file:
|
|||||||
const std = @import("std");
|
const std = @import("std");
|
||||||
const gpu = @import("gpu");
|
const gpu = @import("gpu");
|
||||||
const GpuDevice = gpu.GpuDevice;
|
const GpuDevice = gpu.GpuDevice;
|
||||||
|
const GpuArenaAllocator = gpu.GpuArenaAllocator;
|
||||||
const GpuBuffer = gpu.GpuBuffer;
|
const GpuBuffer = gpu.GpuBuffer;
|
||||||
const GpuRender = gpu.GpuRender;
|
const GpuRender = gpu.GpuRender;
|
||||||
const GpuTexture = gpu.GpuTexture;
|
const GpuTexture = gpu.GpuTexture;
|
||||||
@ -111,19 +116,21 @@ pub fn main(init: std.process.Init) !void {
|
|||||||
const device = try GpuDevice.init(.{});
|
const device = try GpuDevice.init(.{});
|
||||||
defer device.deinit();
|
defer device.deinit();
|
||||||
|
|
||||||
// 2. Get base device GPU Allocator
|
// 2. Init VRAM Arena
|
||||||
const gloc = device.gpuAllocator();
|
var grena = GpuArenaAllocator.init(allocator, device.gpuAllocator());
|
||||||
|
defer grena.deinit();
|
||||||
|
const glloc = grena.gpuAllocator();
|
||||||
|
|
||||||
// 3. Load Render Pipeline
|
// 3. Load Render Pipeline
|
||||||
const circle_rp = try GpuRender.init(
|
const circle_rp = try GpuRender.init(
|
||||||
gloc,
|
glloc,
|
||||||
@embedFile("shaders/circle.wgsl"),
|
@embedFile("shaders/circle.wgsl"),
|
||||||
.{ .bindings = &.{}, .texture_format = .RGBA8Unorm, .topology = .TriangleStrip },
|
.{ .bindings = &.{}, .texture_format = .RGBA8Unorm, .topology = .TriangleStrip },
|
||||||
);
|
);
|
||||||
defer circle_rp.deinit();
|
defer circle_rp.deinit();
|
||||||
|
|
||||||
// 4. Create VRAM texture to render into
|
// 4. Create VRAM texture to render into
|
||||||
const texture = try GpuTexture.init(gloc, .{
|
const texture = try GpuTexture.init(glloc, .{
|
||||||
.format = .RGBA8Unorm,
|
.format = .RGBA8Unorm,
|
||||||
.size = .{ .width = width, .height = height, .depthOrArrayLayers = 1 },
|
.size = .{ .width = width, .height = height, .depthOrArrayLayers = 1 },
|
||||||
.usage = .initMany(&.{ .RenderAttachment, .CopySrc }),
|
.usage = .initMany(&.{ .RenderAttachment, .CopySrc }),
|
||||||
@ -131,21 +138,22 @@ pub fn main(init: std.process.Init) !void {
|
|||||||
defer texture.deinit();
|
defer texture.deinit();
|
||||||
|
|
||||||
// 5. Create a view from texture
|
// 5. Create a view from texture
|
||||||
const view = try GpuTextureView.init(gloc, texture, .{});
|
const view = try GpuTextureView.init(glloc, texture, .{});
|
||||||
defer view.deinit();
|
defer view.deinit();
|
||||||
|
|
||||||
// 6. Run the rendering pipeline
|
// 6. Run the rendering pipeline
|
||||||
try circle_rp.draw(gloc, view, 4, .{});
|
try circle_rp.draw(glloc, view, 4, .{});
|
||||||
|
|
||||||
// 7. Copy Texture into a readable GPU staging buffer
|
// 7. Load Texture into GpuBuffer
|
||||||
const cpu_staging_buf = try texture.buffCopy(gloc);
|
const cpu_staging_cpu = try texture.buffCopy(glloc);
|
||||||
defer cpu_staging_buf.deinit();
|
defer cpu_staging_cpu.deinit();
|
||||||
|
|
||||||
// 8. Read GpuBuffer to CPU memory
|
// 8. Read GpuBuffer to CPU
|
||||||
const pixels = try cpu_staging_buf.read(allocator, u8);
|
// This need to be free manually because CPU memory
|
||||||
|
const pixels = try cpu_staging_cpu.read(allocator, u8);
|
||||||
defer allocator.free(pixels);
|
defer allocator.free(pixels);
|
||||||
|
|
||||||
// 9. Write out to a simple PPM image
|
// 9. Write a simple ppm image
|
||||||
try savePpm(init.io, "circle.ppm", width, height, pixels);
|
try savePpm(init.io, "circle.ppm", width, height, pixels);
|
||||||
}
|
}
|
||||||
|
|
||||||
@ -215,7 +223,7 @@ No extra installation required. Automatically links against `d3d12`, `dxgi`, and
|
|||||||
Add it to your `build.zig.zon`:
|
Add it to your `build.zig.zon`:
|
||||||
|
|
||||||
```bash
|
```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.2.1
|
||||||
```
|
```
|
||||||
|
|
||||||
Then, expose it in your `build.zig`:
|
Then, expose it in your `build.zig`:
|
||||||
|
|||||||
@ -1,6 +1,6 @@
|
|||||||
.{
|
.{
|
||||||
.name = .zig_wgpu,
|
.name = .zig_wgpu,
|
||||||
.version = "0.2.0",
|
.version = "0.2.3",
|
||||||
.fingerprint = 0x5d0e853acbc0c2c6,
|
.fingerprint = 0x5d0e853acbc0c2c6,
|
||||||
.minimum_zig_version = "0.16.0",
|
.minimum_zig_version = "0.16.0",
|
||||||
.dependencies = .{},
|
.dependencies = .{},
|
||||||
|
|||||||
@ -13,10 +13,10 @@ const Vec = struct {
|
|||||||
buf: GpuBuffer,
|
buf: GpuBuffer,
|
||||||
len: usize,
|
len: usize,
|
||||||
|
|
||||||
// Changed: gloc is passed by value (const)
|
// Changed: glloc is passed by value (const)
|
||||||
pub fn initZero(gloc: GpuAllocator, len: usize) !Vec {
|
pub fn initZero(glloc: GpuAllocator, len: usize) !Vec {
|
||||||
return .{
|
return .{
|
||||||
.buf = try GpuBuffer.init(gloc, .{
|
.buf = try GpuBuffer.init(glloc, .{
|
||||||
.size = len * @sizeOf(f16),
|
.size = len * @sizeOf(f16),
|
||||||
.usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }),
|
.usage = .initMany(&.{ .Storage, .CopyDst, .CopySrc }),
|
||||||
}),
|
}),
|
||||||
@ -24,9 +24,9 @@ const Vec = struct {
|
|||||||
};
|
};
|
||||||
}
|
}
|
||||||
|
|
||||||
// Changed: gloc is passed by value
|
// Changed: glloc is passed by value
|
||||||
pub fn initLoad(gloc: GpuAllocator, data: []const f16) !Vec {
|
pub fn initLoad(glloc: GpuAllocator, data: []const f16) !Vec {
|
||||||
var self = try initZero(gloc, data.len);
|
var self = try initZero(glloc, data.len);
|
||||||
try self.load(data); // Direct access via the interface copy
|
try self.load(data); // Direct access via the interface copy
|
||||||
return self;
|
return self;
|
||||||
}
|
}
|
||||||
@ -40,18 +40,18 @@ const Vec = struct {
|
|||||||
try self.buf.load(f16, data);
|
try self.buf.load(f16, data);
|
||||||
}
|
}
|
||||||
|
|
||||||
// Changed: gloc is passed by value instead of *GpuAllocator
|
// Changed: glloc is passed by value instead of *GpuAllocator
|
||||||
pub fn run(self: Vec, gloc: GpuAllocator, other: Vec, process: GpuCompute) !Vec {
|
pub fn run(self: Vec, glloc: GpuAllocator, other: Vec, process: GpuCompute) !Vec {
|
||||||
std.debug.assert(self.len == other.len);
|
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();
|
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;
|
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 {
|
pub fn read(self: Vec, alloc: std.mem.Allocator) ![]f16 {
|
||||||
return self.buf.read(alloc, 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());
|
var grena = GpuArenaAllocator.init(init.gpa, device.gpuAllocator());
|
||||||
defer grena.deinit();
|
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) },
|
.{ .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 ---
|
// --- 1. GPU ALLOCATION PHASE ---
|
||||||
const alloc_start = std.Io.Clock.awake.now(init.io);
|
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();
|
defer a.deinit();
|
||||||
const b = try Vec.initLoad(gloc, data_b);
|
const b = try Vec.initLoad(glloc, data_b);
|
||||||
defer b.deinit();
|
defer b.deinit();
|
||||||
|
|
||||||
const alloc_duration = alloc_start.durationTo(std.Io.Clock.awake.now(init.io));
|
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 ---
|
// --- 2. COMPUTE PHASE ---
|
||||||
const compute_start = std.Io.Clock.awake.now(init.io);
|
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();
|
defer sum.deinit();
|
||||||
|
|
||||||
// All 3 buffers (a, b, sum) are currently resident in VRAM here.
|
// 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
|
// 2. Init VRAM Arena
|
||||||
var grena = GpuArenaAllocator.init(allocator, device.gpuAllocator());
|
var grena = GpuArenaAllocator.init(allocator, device.gpuAllocator());
|
||||||
defer grena.deinit();
|
defer grena.deinit();
|
||||||
const gloc = grena.gpuAllocator();
|
const glloc = grena.gpuAllocator();
|
||||||
|
|
||||||
// 3. Load Render Pipeline
|
// 3. Load Render Pipeline
|
||||||
const circle_rp = try GpuRender.init(
|
const circle_rp = try GpuRender.init(
|
||||||
gloc,
|
glloc,
|
||||||
@embedFile("shaders/circle.wgsl"),
|
@embedFile("shaders/circle.wgsl"),
|
||||||
.{ .bindings = &.{}, .texture_format = .RGBA8Unorm, .topology = .TriangleStrip },
|
.{ .bindings = &.{}, .texture_format = .RGBA8Unorm, .topology = .TriangleStrip },
|
||||||
);
|
);
|
||||||
defer circle_rp.deinit();
|
defer circle_rp.deinit();
|
||||||
|
|
||||||
// 4. Create VRAM texture to render into
|
// 4. Create VRAM texture to render into
|
||||||
const texture = try GpuTexture.init(gloc, .{
|
const texture = try GpuTexture.init(glloc, .{
|
||||||
.format = .RGBA8Unorm,
|
.format = .RGBA8Unorm,
|
||||||
.size = .{ .width = width, .height = height, .depthOrArrayLayers = 1 },
|
.size = .{ .width = width, .height = height, .depthOrArrayLayers = 1 },
|
||||||
.usage = .initMany(&.{ .RenderAttachment, .CopySrc }),
|
.usage = .initMany(&.{ .RenderAttachment, .CopySrc }),
|
||||||
@ -39,14 +39,14 @@ pub fn main(init: std.process.Init) !void {
|
|||||||
defer texture.deinit();
|
defer texture.deinit();
|
||||||
|
|
||||||
// 5. Create a view from texture
|
// 5. Create a view from texture
|
||||||
const view = try GpuTextureView.init(gloc, texture, .{});
|
const view = try GpuTextureView.init(glloc, texture, .{});
|
||||||
defer view.deinit();
|
defer view.deinit();
|
||||||
|
|
||||||
// 6. Run the rendering pipeline
|
// 6. Run the rendering pipeline
|
||||||
try circle_rp.draw(gloc, view, 4, .{});
|
try circle_rp.draw(glloc, view, 4, .{});
|
||||||
|
|
||||||
// 7. Load Texture into GpuBuffer
|
// 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();
|
defer cpu_staging_cpu.deinit();
|
||||||
|
|
||||||
// 8. Read GpuBuffer to CPU
|
// 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
|
// 2. Create a GPU Arena to manage VRAM
|
||||||
var grena = GpuArenaAllocator.init(allocator, device.gpuAllocator());
|
var grena = GpuArenaAllocator.init(allocator, device.gpuAllocator());
|
||||||
defer grena.deinit();
|
defer grena.deinit();
|
||||||
const gloc = grena.gpuAllocator();
|
const glloc = grena.gpuAllocator();
|
||||||
|
|
||||||
// 3. Load the WGSL compute pipeline
|
// 3. Load the WGSL compute pipeline
|
||||||
const add_cp = try GpuCompute.init(
|
const add_cp = try GpuCompute.init(
|
||||||
gloc,
|
glloc,
|
||||||
@embedFile("shaders/add.wgsl"),
|
@embedFile("shaders/add.wgsl"),
|
||||||
.{
|
.{
|
||||||
.label = "add",
|
.label = "add",
|
||||||
@ -45,9 +45,9 @@ pub fn main(init: std.process.Init) !void {
|
|||||||
|
|
||||||
// 5. Initialize raw GPU Buffers
|
// 5. Initialize raw GPU Buffers
|
||||||
const byte_size = len * @sizeOf(f16);
|
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_a = try GpuBuffer.init(glloc, .{ .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_b = try GpuBuffer.init(glloc, .{ .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_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
|
// 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.
|
// release them at the end. You can also manually call buf_x.deinit() if desired.
|
||||||
@ -58,7 +58,7 @@ pub fn main(init: std.process.Init) !void {
|
|||||||
try buf_b.load(f16, data_b);
|
try buf_b.load(f16, data_b);
|
||||||
|
|
||||||
// 7. Dispatch the Compute
|
// 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
|
// 8. Map and copy the resulting buffer back to the CPU
|
||||||
const out = try buf_out.read(allocator, f16);
|
const out = try buf_out.read(allocator, f16);
|
||||||
|
|||||||
@ -6,7 +6,7 @@ const svOpt = @import("utils.zig").svOpt;
|
|||||||
raw: c.WGPUBuffer,
|
raw: c.WGPUBuffer,
|
||||||
size: u64,
|
size: u64,
|
||||||
usage: c.WGPUBufferUsage,
|
usage: c.WGPUBufferUsage,
|
||||||
gloc: GpuAllocator,
|
glloc: GpuAllocator,
|
||||||
|
|
||||||
const BufferUsage = enum(u64) {
|
const BufferUsage = enum(u64) {
|
||||||
None = 0x0000000000000000,
|
None = 0x0000000000000000,
|
||||||
@ -28,7 +28,7 @@ const GpuBufferDef = struct {
|
|||||||
usage: std.EnumSet(BufferUsage),
|
usage: std.EnumSet(BufferUsage),
|
||||||
};
|
};
|
||||||
|
|
||||||
pub fn init(gloc: GpuAllocator, def: GpuBufferDef) !@This() {
|
pub fn init(glloc: GpuAllocator, def: GpuBufferDef) !@This() {
|
||||||
var use: u64 = 0;
|
var use: u64 = 0;
|
||||||
var iter = def.usage.iterator();
|
var iter = def.usage.iterator();
|
||||||
while (iter.next()) |flag| use |= @intFromEnum(flag);
|
while (iter.next()) |flag| use |= @intFromEnum(flag);
|
||||||
@ -36,7 +36,7 @@ pub fn init(gloc: GpuAllocator, def: GpuBufferDef) !@This() {
|
|||||||
// Automatically align the buffer size forward to a multiple of 4 bytes under the hood
|
// 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 aligned_size = std.mem.alignForward(u64, def.size, 4);
|
||||||
|
|
||||||
const raw_handle = try gloc.allocBuffer(.{
|
const raw_handle = try glloc.allocBuffer(.{
|
||||||
.size = aligned_size,
|
.size = aligned_size,
|
||||||
.usage = use,
|
.usage = use,
|
||||||
.label = svOpt(def.label),
|
.label = svOpt(def.label),
|
||||||
@ -45,12 +45,12 @@ pub fn init(gloc: GpuAllocator, def: GpuBufferDef) !@This() {
|
|||||||
.raw = raw_handle,
|
.raw = raw_handle,
|
||||||
.size = aligned_size,
|
.size = aligned_size,
|
||||||
.usage = use,
|
.usage = use,
|
||||||
.gloc = gloc,
|
.glloc = glloc,
|
||||||
};
|
};
|
||||||
}
|
}
|
||||||
|
|
||||||
pub fn deinit(self: @This()) void {
|
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 {
|
pub fn getConstMappedRange(self: @This(), offset: u64, size: u64) ?*const anyopaque {
|
||||||
@ -81,20 +81,20 @@ pub fn load(
|
|||||||
|
|
||||||
if (bytes == self.size) {
|
if (bytes == self.size) {
|
||||||
// Aligned path: direct download
|
// Aligned path: direct download
|
||||||
c.wgpuQueueWriteBuffer(self.gloc.device.queue, self.raw, 0, data.ptr, self.size);
|
c.wgpuQueueWriteBuffer(self.glloc.device.queue, self.raw, 0, data.ptr, self.size);
|
||||||
} else {
|
} else {
|
||||||
// Unaligned path: Split the write into an aligned chunk and a padded remainder
|
// Unaligned path: Split the write into an aligned chunk and a padded remainder
|
||||||
// to support arbitrary lengths without any allocations or large stack arrays.
|
// to support arbitrary lengths without any allocations or large stack arrays.
|
||||||
const aligned_part = (bytes / 4) * 4;
|
const aligned_part = (bytes / 4) * 4;
|
||||||
if (aligned_part > 0) {
|
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 };
|
var remainder_buf: [4]u8 = .{ 0, 0, 0, 0 };
|
||||||
const data_bytes = std.mem.sliceAsBytes(data);
|
const data_bytes = std.mem.sliceAsBytes(data);
|
||||||
@memcpy(remainder_buf[0 .. bytes - aligned_part], data_bytes[aligned_part..bytes]);
|
@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);
|
||||||
}
|
}
|
||||||
}
|
}
|
||||||
|
|
||||||
@ -102,19 +102,19 @@ pub fn load(
|
|||||||
pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
||||||
const out = try alloc.alloc(T, @divExact(self.size, @sizeOf(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,
|
.size = self.size,
|
||||||
.usage = .initMany(&.{ .MapRead, .CopyDst }),
|
.usage = .initMany(&.{ .MapRead, .CopyDst }),
|
||||||
.label = "staging_read_buffer",
|
.label = "staging_read_buffer",
|
||||||
});
|
});
|
||||||
defer staging.deinit();
|
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;
|
||||||
c.wgpuCommandEncoderCopyBufferToBuffer(enc, self.raw, 0, staging.raw, 0, self.size);
|
c.wgpuCommandEncoderCopyBufferToBuffer(enc, self.raw, 0, staging.raw, 0, self.size);
|
||||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||||
defer c.wgpuCommandEncoderRelease(enc);
|
defer c.wgpuCommandEncoderRelease(enc);
|
||||||
defer c.wgpuCommandBufferRelease(cmd);
|
defer c.wgpuCommandBufferRelease(cmd);
|
||||||
c.wgpuQueueSubmit(self.gloc.device.queue, 1, &cmd);
|
c.wgpuQueueSubmit(self.glloc.device.queue, 1, &cmd);
|
||||||
|
|
||||||
var mapped = false;
|
var mapped = false;
|
||||||
staging.mapAsync(
|
staging.mapAsync(
|
||||||
@ -123,7 +123,7 @@ pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
|||||||
self.size,
|
self.size,
|
||||||
.{ .callback = onMapped, .userdata1 = &mapped },
|
.{ .callback = onMapped, .userdata1 = &mapped },
|
||||||
);
|
);
|
||||||
while (!mapped) self.gloc.device.poll();
|
while (!mapped) self.glloc.device.poll();
|
||||||
|
|
||||||
const ptr: [*]const T = @ptrCast(@alignCast(
|
const ptr: [*]const T = @ptrCast(@alignCast(
|
||||||
staging.getConstMappedRange(0, self.size),
|
staging.getConstMappedRange(0, self.size),
|
||||||
|
|||||||
@ -22,40 +22,40 @@ pub const ComputeDef = struct {
|
|||||||
};
|
};
|
||||||
|
|
||||||
pip: c.WGPUComputePipeline,
|
pip: c.WGPUComputePipeline,
|
||||||
gloc: GpuAllocator,
|
glloc: GpuAllocator,
|
||||||
def: ComputeDef,
|
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{
|
var wgsl_src = c.WGPUShaderSourceWGSL{
|
||||||
.chain = .{ .sType = c.WGPUSType_ShaderSourceWGSL },
|
.chain = .{ .sType = c.WGPUSType_ShaderSourceWGSL },
|
||||||
.code = sv(wgsl),
|
.code = sv(wgsl),
|
||||||
};
|
};
|
||||||
const shader = c.wgpuDeviceCreateShaderModule(gloc.device.device, &.{
|
const shader = c.wgpuDeviceCreateShaderModule(glloc.device.device, &.{
|
||||||
.nextInChain = @ptrCast(&wgsl_src),
|
.nextInChain = @ptrCast(&wgsl_src),
|
||||||
}) orelse return error.Shader;
|
}) orelse return error.Shader;
|
||||||
defer c.wgpuShaderModuleRelease(shader);
|
defer c.wgpuShaderModuleRelease(shader);
|
||||||
|
|
||||||
const pip = try gloc.allocComputePipeline(.{
|
const pip = try glloc.allocComputePipeline(.{
|
||||||
.label = svOpt(def.label),
|
.label = svOpt(def.label),
|
||||||
.compute = .{ .module = shader, .entryPoint = sv("main") },
|
.compute = .{ .module = shader, .entryPoint = sv("main") },
|
||||||
});
|
});
|
||||||
|
|
||||||
return .{
|
return .{
|
||||||
.gloc = gloc,
|
.glloc = glloc,
|
||||||
.pip = pip,
|
.pip = pip,
|
||||||
.def = def,
|
.def = def,
|
||||||
};
|
};
|
||||||
}
|
}
|
||||||
|
|
||||||
pub fn deinit(self: @This()) void {
|
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.
|
/// 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(
|
pub fn run(
|
||||||
self: @This(),
|
self: @This(),
|
||||||
gloc: GpuAllocator,
|
glloc: GpuAllocator,
|
||||||
args: anytype,
|
args: anytype,
|
||||||
) !void {
|
) !void {
|
||||||
const type_info = @typeInfo(@TypeOf(args));
|
const type_info = @typeInfo(@TypeOf(args));
|
||||||
@ -113,12 +113,12 @@ pub fn run(
|
|||||||
defer if (info_buf) |b| b.deinit();
|
defer if (info_buf) |b| b.deinit();
|
||||||
|
|
||||||
if (self.def.append_info_buffer) {
|
if (self.def.append_info_buffer) {
|
||||||
info_buf = try GpuBuffer.init(gloc, .{
|
info_buf = try GpuBuffer.init(glloc, .{
|
||||||
.size = @sizeOf(u32),
|
.size = @sizeOf(u32),
|
||||||
.usage = .initMany(&.{ .Uniform, .CopyDst }),
|
.usage = .initMany(&.{ .Uniform, .CopyDst }),
|
||||||
.label = "compute_info_buffer",
|
.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] = .{
|
entries_buf[entry_count] = .{
|
||||||
.binding = @intCast(entry_count),
|
.binding = @intCast(entry_count),
|
||||||
@ -130,11 +130,11 @@ pub fn run(
|
|||||||
}
|
}
|
||||||
|
|
||||||
const entries = entries_buf[0..entry_count];
|
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(
|
fn submitPass(
|
||||||
gloc: GpuAllocator,
|
glloc: GpuAllocator,
|
||||||
pipeline: c.WGPUComputePipeline,
|
pipeline: c.WGPUComputePipeline,
|
||||||
entries: []const c.WGPUBindGroupEntry,
|
entries: []const c.WGPUBindGroupEntry,
|
||||||
n: usize,
|
n: usize,
|
||||||
@ -146,14 +146,14 @@ fn submitPass(
|
|||||||
const bgl = c.wgpuComputePipelineGetBindGroupLayout(pipeline, 0);
|
const bgl = c.wgpuComputePipelineGetBindGroupLayout(pipeline, 0);
|
||||||
defer c.wgpuBindGroupLayoutRelease(bgl);
|
defer c.wgpuBindGroupLayoutRelease(bgl);
|
||||||
|
|
||||||
const bg = c.wgpuDeviceCreateBindGroup(gloc.device.device, &.{
|
const bg = c.wgpuDeviceCreateBindGroup(glloc.device.device, &.{
|
||||||
.layout = bgl,
|
.layout = bgl,
|
||||||
.entries = entries.ptr,
|
.entries = entries.ptr,
|
||||||
.entryCount = entries.len,
|
.entryCount = entries.len,
|
||||||
}) orelse return error.BindGroup;
|
}) orelse return error.BindGroup;
|
||||||
defer c.wgpuBindGroupRelease(bg);
|
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);
|
const pass = c.wgpuCommandEncoderBeginComputePass(enc, null);
|
||||||
c.wgpuComputePassEncoderSetPipeline(pass, pipeline);
|
c.wgpuComputePassEncoderSetPipeline(pass, pipeline);
|
||||||
c.wgpuComputePassEncoderSetBindGroup(pass, 0, bg, 0, null);
|
c.wgpuComputePassEncoderSetBindGroup(pass, 0, bg, 0, null);
|
||||||
@ -168,7 +168,7 @@ fn submitPass(
|
|||||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||||
defer c.wgpuCommandEncoderRelease(enc);
|
defer c.wgpuCommandEncoderRelease(enc);
|
||||||
defer c.wgpuCommandBufferRelease(cmd);
|
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 {
|
fn ceilDiv(n: usize, d: usize) usize {
|
||||||
|
|||||||
@ -34,16 +34,16 @@ const GpuPrimitiveTopology = enum(c_uint) {
|
|||||||
Force32 = 0x7FFFFFFF,
|
Force32 = 0x7FFFFFFF,
|
||||||
};
|
};
|
||||||
|
|
||||||
gloc: GpuAllocator,
|
glloc: GpuAllocator,
|
||||||
pip: c.WGPURenderPipeline,
|
pip: c.WGPURenderPipeline,
|
||||||
def: GpuRenderDef,
|
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{
|
var wgsl_src = c.WGPUShaderSourceWGSL{
|
||||||
.chain = .{ .sType = c.WGPUSType_ShaderSourceWGSL },
|
.chain = .{ .sType = c.WGPUSType_ShaderSourceWGSL },
|
||||||
.code = sv(wgsl),
|
.code = sv(wgsl),
|
||||||
};
|
};
|
||||||
const shader = c.wgpuDeviceCreateShaderModule(gloc.device.device, &.{
|
const shader = c.wgpuDeviceCreateShaderModule(glloc.device.device, &.{
|
||||||
.nextInChain = @ptrCast(&wgsl_src),
|
.nextInChain = @ptrCast(&wgsl_src),
|
||||||
}) orelse return error.Shader;
|
}) orelse return error.Shader;
|
||||||
defer c.wgpuShaderModuleRelease(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
|
// 3. Compile the Complete Render Pipeline
|
||||||
const pip = try gloc.allocRenderPipeline(.{
|
const pip = try glloc.allocRenderPipeline(.{
|
||||||
.label = svOpt(def.label),
|
.label = svOpt(def.label),
|
||||||
.vertex = .{
|
.vertex = .{
|
||||||
.module = shader,
|
.module = shader,
|
||||||
@ -90,21 +90,21 @@ pub fn init(gloc: GpuAllocator, wgsl: []const u8, def: GpuRenderDef) !@This() {
|
|||||||
});
|
});
|
||||||
|
|
||||||
return .{
|
return .{
|
||||||
.gloc = gloc,
|
.glloc = glloc,
|
||||||
.pip = pip,
|
.pip = pip,
|
||||||
.def = def,
|
.def = def,
|
||||||
};
|
};
|
||||||
}
|
}
|
||||||
|
|
||||||
pub fn deinit(self: @This()) void {
|
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.
|
/// Execute the render pass targeting a specific frame texture view.
|
||||||
/// Passes bind groups via a tuple exactly like your original compute setup.
|
/// Passes bind groups via a tuple exactly like your original compute setup.
|
||||||
pub fn draw(
|
pub fn draw(
|
||||||
self: @This(),
|
self: @This(),
|
||||||
gloc: GpuAllocator,
|
glloc: GpuAllocator,
|
||||||
target_view: GpuTextureView,
|
target_view: GpuTextureView,
|
||||||
vertex_count: u32,
|
vertex_count: u32,
|
||||||
args: anytype,
|
args: anytype,
|
||||||
@ -138,7 +138,7 @@ pub fn draw(
|
|||||||
const bgl = c.wgpuRenderPipelineGetBindGroupLayout(self.pip, 0);
|
const bgl = c.wgpuRenderPipelineGetBindGroupLayout(self.pip, 0);
|
||||||
defer c.wgpuBindGroupLayoutRelease(bgl);
|
defer c.wgpuBindGroupLayoutRelease(bgl);
|
||||||
|
|
||||||
const bg = c.wgpuDeviceCreateBindGroup(gloc.device.device, &.{
|
const bg = c.wgpuDeviceCreateBindGroup(glloc.device.device, &.{
|
||||||
.layout = bgl,
|
.layout = bgl,
|
||||||
.entries = entries.ptr,
|
.entries = entries.ptr,
|
||||||
.entryCount = @intCast(entries.len),
|
.entryCount = @intCast(entries.len),
|
||||||
@ -146,7 +146,7 @@ pub fn draw(
|
|||||||
defer c.wgpuBindGroupRelease(bg);
|
defer c.wgpuBindGroupRelease(bg);
|
||||||
|
|
||||||
// Encode Render Command
|
// 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);
|
defer c.wgpuCommandEncoderRelease(enc);
|
||||||
|
|
||||||
const color_attachment = c.WGPURenderPassColorAttachment{
|
const color_attachment = c.WGPURenderPassColorAttachment{
|
||||||
@ -180,5 +180,5 @@ pub fn draw(
|
|||||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||||
defer c.wgpuCommandBufferRelease(cmd);
|
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,
|
raw: c.WGPUTexture,
|
||||||
gloc: GpuAllocator,
|
glloc: GpuAllocator,
|
||||||
def: GpuTextureDef,
|
def: GpuTextureDef,
|
||||||
|
|
||||||
pub fn init(gloc: GpuAllocator, def: GpuTextureDef) !@This() {
|
pub fn init(glloc: GpuAllocator, def: GpuTextureDef) !@This() {
|
||||||
var use: u64 = 0;
|
var use: u64 = 0;
|
||||||
var iter = def.usage.iterator();
|
var iter = def.usage.iterator();
|
||||||
while (iter.next()) |flag| use |= @intFromEnum(flag);
|
while (iter.next()) |flag| use |= @intFromEnum(flag);
|
||||||
@ -31,13 +31,13 @@ pub fn init(gloc: GpuAllocator, def: GpuTextureDef) !@This() {
|
|||||||
.mipLevelCount = 1,
|
.mipLevelCount = 1,
|
||||||
.sampleCount = 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 {
|
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 {
|
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.
|
/// Return a GpuBuffer containing a copy of the texture.
|
||||||
pub fn buffCopy(self: @This(), gloc: GpuAllocator) !GpuBuffer {
|
pub fn buffCopy(self: @This(), glloc: GpuAllocator) !GpuBuffer {
|
||||||
const buf = try GpuBuffer.init(gloc, .{
|
const buf = try GpuBuffer.init(glloc, .{
|
||||||
.size = self.bytesSize(),
|
.size = self.bytesSize(),
|
||||||
.usage = .initMany(&.{ .CopyDst, .CopySrc }),
|
.usage = .initMany(&.{ .CopyDst, .CopySrc }),
|
||||||
.label = "texture_copy_buffer",
|
.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);
|
defer c.wgpuCommandEncoderRelease(enc);
|
||||||
|
|
||||||
const src_copy = c.WGPUTexelCopyTextureInfo{
|
const src_copy = c.WGPUTexelCopyTextureInfo{
|
||||||
@ -82,7 +82,7 @@ pub fn buffCopy(self: @This(), gloc: GpuAllocator) !GpuBuffer {
|
|||||||
|
|
||||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||||
defer c.wgpuCommandBufferRelease(cmd);
|
defer c.wgpuCommandBufferRelease(cmd);
|
||||||
c.wgpuQueueSubmit(gloc.device.queue, 1, &cmd);
|
c.wgpuQueueSubmit(glloc.device.queue, 1, &cmd);
|
||||||
|
|
||||||
return buf;
|
return buf;
|
||||||
}
|
}
|
||||||
@ -110,7 +110,7 @@ pub fn load(
|
|||||||
const bytes = data.len * @sizeOf(T);
|
const bytes = data.len * @sizeOf(T);
|
||||||
|
|
||||||
c.wgpuQueueWriteTexture(
|
c.wgpuQueueWriteTexture(
|
||||||
self.gloc.device.queue,
|
self.glloc.device.queue,
|
||||||
&.{
|
&.{
|
||||||
.texture = self.raw,
|
.texture = self.raw,
|
||||||
.mipLevel = 0,
|
.mipLevel = 0,
|
||||||
@ -132,14 +132,14 @@ pub fn load(
|
|||||||
pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
||||||
const out = try alloc.alloc(T, @divExact(self.size, @sizeOf(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,
|
.size = self.size,
|
||||||
.usage = .initMany(&.{ .MapRead, .CopyDst }),
|
.usage = .initMany(&.{ .MapRead, .CopyDst }),
|
||||||
.label = "texture_read_staging",
|
.label = "texture_read_staging",
|
||||||
});
|
});
|
||||||
defer staging.deinit();
|
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{
|
const src_copy = c.WGPUTexelCopyTextureInfo{
|
||||||
.texture = self.raw,
|
.texture = self.raw,
|
||||||
.mipLevel = 0,
|
.mipLevel = 0,
|
||||||
@ -158,7 +158,7 @@ pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
|||||||
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
const cmd = c.wgpuCommandEncoderFinish(enc, null);
|
||||||
defer c.wgpuCommandEncoderRelease(enc);
|
defer c.wgpuCommandEncoderRelease(enc);
|
||||||
defer c.wgpuCommandBufferRelease(cmd);
|
defer c.wgpuCommandBufferRelease(cmd);
|
||||||
c.wgpuQueueSubmit(self.gloc.device.queue, 1, &cmd);
|
c.wgpuQueueSubmit(self.glloc.device.queue, 1, &cmd);
|
||||||
|
|
||||||
var mapped = false;
|
var mapped = false;
|
||||||
staging.mapAsync(
|
staging.mapAsync(
|
||||||
@ -167,7 +167,7 @@ pub fn read(self: @This(), alloc: std.mem.Allocator, T: type) ![]T {
|
|||||||
self.size,
|
self.size,
|
||||||
.{ .callback = onMapped, .userdata1 = &mapped },
|
.{ .callback = onMapped, .userdata1 = &mapped },
|
||||||
);
|
);
|
||||||
while (!mapped) self.gloc.device.poll();
|
while (!mapped) self.glloc.device.poll();
|
||||||
|
|
||||||
const ptr: [*]const T = @ptrCast(@alignCast(
|
const ptr: [*]const T = @ptrCast(@alignCast(
|
||||||
staging.getConstMappedRange(0, self.size),
|
staging.getConstMappedRange(0, self.size),
|
||||||
|
|||||||
@ -13,23 +13,23 @@ pub const GpuViewDef = struct {
|
|||||||
};
|
};
|
||||||
|
|
||||||
raw: c.WGPUTextureView,
|
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 use: u64 = 0;
|
||||||
var iter = def.usage.iterator();
|
var iter = def.usage.iterator();
|
||||||
while (iter.next()) |flag| use |= @intFromEnum(flag);
|
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),
|
.label = svOpt(def.label),
|
||||||
.format = @intFromEnum(def.format),
|
.format = @intFromEnum(def.format),
|
||||||
.usage = use,
|
.usage = use,
|
||||||
.mipLevelCount = 1,
|
.mipLevelCount = 1,
|
||||||
.arrayLayerCount = 1,
|
.arrayLayerCount = 1,
|
||||||
});
|
});
|
||||||
return .{ .gloc = gloc, .raw = raw };
|
return .{ .glloc = glloc, .raw = raw };
|
||||||
}
|
}
|
||||||
|
|
||||||
pub fn deinit(self: @This()) void {
|
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