Skip to content

Commit

Permalink
Remove vertex_pulling_transfrom from PipelineCompilationOptions.
Browse files Browse the repository at this point in the history 
This option was only evaluated for Metal backends, and now it's required
there so the option is going away. It is still configurable for tests
via the PipelineOptions struct, deserialized from .ron files.

This also fixes some type problems with the unpack functions in
writer.rs. Metal << operator extends operand to int-sized, which then
has to be cast back down to the real size before as_type bit conversion.
The math for the snorm values is corrected, in some cases using the
metal unpack_snorm2x16_to_float function because we can't directly
cast a bit-shifted ushort value to half.
  • Loading branch information
@bradwerth
bradwerth committed Jul 15, 2024
1 parent f44f52a commit 2bf3d26
Show file tree
Hide file tree
Showing 17 changed files with 764 additions and 104 deletions.
3 changes: 0 additions & 3 deletions deno_webgpu/pipeline.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -114,7 +114,6 @@ pub fn op_webgpu_create_compute_pipeline(
entry_point: compute.entry_point.map(Cow::from),
constants: Cow::Owned(compute.constants.unwrap_or_default()),
zero_initialize_workgroup_memory: true,
vertex_pulling_transform: false,
},
cache: None,
};
Expand Down Expand Up @@ -359,7 +358,6 @@ pub fn op_webgpu_create_render_pipeline(
constants: Cow::Owned(fragment.constants.unwrap_or_default()),
// Required to be true for WebGPU
zero_initialize_workgroup_memory: true,
vertex_pulling_transform: false,
},
targets: Cow::Owned(fragment.targets),
})
Expand All @@ -385,7 +383,6 @@ pub fn op_webgpu_create_render_pipeline(
constants: Cow::Owned(args.vertex.constants.unwrap_or_default()),
// Required to be true for WebGPU
zero_initialize_workgroup_memory: true,
vertex_pulling_transform: false,
},
buffers: Cow::Owned(vertex_buffers),
},
Expand Down
1 change: 1 addition & 0 deletions naga/CHANGELOG.md
View file
Original file line number Diff line number Diff line change
Expand Up @@ -81,6 +81,7 @@ For changelogs after v0.14, see [the wgpu changelog](../CHANGELOG.md).
- Make varyings' struct members unique. ([#2521](https://github.com/gfx-rs/naga/pull/2521)) **@evahop**
- Add experimental vertex pulling transform flag. ([#5254](https://github.com/gfx-rs/wgpu/pull/5254)) **@bradwerth**
- Fixup some generated MSL for vertex buffer unpack functions. ([#5829](https://github.com/gfx-rs/wgpu/pull/5829)) **@bradwerth**
- Make vertex pulling transform on by default. ([#5773](https://github.com/gfx-rs/wgpu/pull/5773)) **@bradwerth**

#### GLSL-OUT

Expand Down
4 changes: 3 additions & 1 deletion naga/src/back/msl/mod.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -354,7 +354,9 @@ pub struct PipelineOptions {
/// to receive the vertex buffers, lengths, and vertex id as args,
/// and bounds-check the vertex id and use the index into the
/// vertex buffers to access attributes, rather than using Metal's
/// [[stage-in]] assembled attribute data.
/// [[stage-in]] assembled attribute data. This is true by default,
/// but remains configurable for use by tests via deserialization
/// of this struct. There is no user-facing way to set this value.
pub vertex_pulling_transform: bool,

/// vertex_buffer_mappings are used during shader translation to
Expand Down
66 changes: 37 additions & 29 deletions naga/src/back/msl/writer.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -3953,8 +3953,8 @@ impl<W: Write> Writer<W> {
)?;
writeln!(
self.out,
"{}return metal::float2((float(b0) - 128.0f) / 255.0f, \
(float(b1) - 128.0f) / 255.0f);",
"{}return metal::float2(metal::max(-1.0f, as_type<char>(b0) / 127.0f), \
metal::max(-1.0f, as_type<char>(b1) / 127.0f));",
back::INDENT
)?;
writeln!(self.out, "}}")?;
Expand All @@ -3971,10 +3971,10 @@ impl<W: Write> Writer<W> {
)?;
writeln!(
self.out,
"{}return metal::float4((float(b0) - 128.0f) / 255.0f, \
(float(b1) - 128.0f) / 255.0f, \
(float(b2) - 128.0f) / 255.0f, \
(float(b3) - 128.0f) / 255.0f);",
"{}return metal::float4(metal::max(-1.0f, as_type<char>(b0) / 127.0f), \
metal::max(-1.0f, as_type<char>(b1) / 127.0f), \
metal::max(-1.0f, as_type<char>(b2) / 127.0f), \
metal::max(-1.0f, as_type<char>(b3) / 127.0f));",
back::INDENT
)?;
writeln!(self.out, "}}")?;
Expand Down Expand Up @@ -4033,8 +4033,8 @@ impl<W: Write> Writer<W> {
)?;
writeln!(
self.out,
"{}return metal::int2(as_type<metal::short>(b1 << 8 | b0), \
as_type<metal::short>(b3 << 8 | b2));",
"{}return metal::int2(as_type<short>(metal::ushort(b1 << 8 | b0)), \
as_type<short>(metal::ushort(b3 << 8 | b2)));",
back::INDENT
)?;
writeln!(self.out, "}}")?;
Expand All @@ -4055,10 +4055,10 @@ impl<W: Write> Writer<W> {
)?;
writeln!(
self.out,
"{}return metal::int4(as_type<metal::short>(b1 << 8 | b0), \
as_type<metal::short>(b3 << 8 | b2), \
as_type<metal::short>(b5 << 8 | b4), \
as_type<metal::short>(b7 << 8 | b6));",
"{}return metal::int4(as_type<short>(metal::ushort(b1 << 8 | b0)), \
as_type<short>(metal::ushort(b3 << 8 | b2)), \
as_type<short>(metal::ushort(b5 << 8 | b4)), \
as_type<short>(metal::ushort(b7 << 8 | b6)));",
back::INDENT
)?;
writeln!(self.out, "}}")?;
Expand Down Expand Up @@ -4117,8 +4117,7 @@ impl<W: Write> Writer<W> {
)?;
writeln!(
self.out,
"{}return metal::float2((float(b1 << 8 | b0) - 32767.0f) / 65535.0f, \
(float(b3 << 8 | b2) - 32767.0f) / 65535.0f);",
"{}return metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2);",
back::INDENT
)?;
writeln!(self.out, "}}")?;
Expand All @@ -4139,10 +4138,8 @@ impl<W: Write> Writer<W> {
)?;
writeln!(
self.out,
"{}return metal::float4((float(b1 << 8 | b0) - 32767.0f) / 65535.0f, \
(float(b3 << 8 | b2) - 32767.0f) / 65535.0f, \
(float(b5 << 8 | b4) - 32767.0f) / 65535.0f, \
(float(b7 << 8 | b6) - 32767.0f) / 65535.0f);",
"{}return metal::float4(metal::unpack_snorm2x16_to_float(b1 << 24 | b0 << 16 | b3 << 8 | b2), \
metal::unpack_snorm2x16_to_float(b5 << 24 | b4 << 16 | b7 << 8 | b6));",
back::INDENT
)?;
writeln!(self.out, "}}")?;
Expand All @@ -4159,8 +4156,8 @@ impl<W: Write> Writer<W> {
)?;
writeln!(
self.out,
"{}return metal::float2(as_type<metal::half>(b1 << 8 | b0), \
as_type<metal::half>(b3 << 8 | b2));",
"{}return metal::float2(as_type<half>(metal::ushort(b1 << 8 | b0)), \
as_type<half>(metal::ushort(b3 << 8 | b2)));",
back::INDENT
)?;
writeln!(self.out, "}}")?;
Expand All @@ -4181,10 +4178,10 @@ impl<W: Write> Writer<W> {
)?;
writeln!(
self.out,
"{}return metal::int4(as_type<metal::half>(b1 << 8 | b0), \
as_type<metal::half>(b3 << 8 | b2), \
as_type<metal::half>(b5 << 8 | b4), \
as_type<metal::half>(b7 << 8 | b6));",
"{}return metal::int4(as_type<half>(metal::ushort(b1 << 8 | b0)), \
as_type<half>(metal::ushort(b3 << 8 | b2)), \
as_type<half>(metal::ushort(b5 << 8 | b4)), \
as_type<half>(metal::ushort(b7 << 8 | b6)));",
back::INDENT
)?;
writeln!(self.out, "}}")?;
Expand Down Expand Up @@ -4390,10 +4387,10 @@ impl<W: Write> Writer<W> {
let name = self.namer.call("unpackSint32");
writeln!(
self.out,
"metal::int {name}(uint b0, \
uint b1, \
uint b2, \
uint b3) {{"
"int {name}(uint b0, \
uint b1, \
uint b2, \
uint b3) {{"
)?;
writeln!(
self.out,
Expand Down Expand Up @@ -4495,7 +4492,18 @@ impl<W: Write> Writer<W> {
)?;
writeln!(
self.out,
"{}return unpack_unorm10a2_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0);",
// The following is correct for RGBA packing, but our format seems to
// match ABGR, which can be fed into the Metal builtin function
// unpack_unorm10a2_to_float.
/*
"{}uint v = (b3 << 24 | b2 << 16 | b1 << 8 | b0); \
uint r = (v & 0xFFC00000) >> 22; \
uint g = (v & 0x003FF000) >> 12; \
uint b = (v & 0x00000FFC) >> 2; \
uint a = (v & 0x00000003); \
return metal::float4(float(r) / 1023.0f, float(g) / 1023.0f, float(b) / 1023.0f, float(a) / 3.0f);",
*/
"{}return metal::unpack_unorm10a2_to_float(b3 << 24 | b2 << 16 | b1 << 8 | b0);",
back::INDENT
)?;
writeln!(self.out, "}}")?;
Expand Down
1 change: 1 addition & 0 deletions tests/tests/root.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -42,6 +42,7 @@ mod subgroup_operations;
mod texture_bounds;
mod texture_view_creation;
mod transfer;
mod vertex_formats;
mod vertex_indices;
mod write_texture;
mod zero_init_texture_after_discard;
Expand Down
Loading

0 comments on commit 2bf3d26

Please sign in to comment.