diff --git a/CHANGELOG.md b/CHANGELOG.md index 4e42655189b..1cc2d3a985a 100644 --- a/CHANGELOG.md +++ b/CHANGELOG.md @@ -78,6 +78,12 @@ SamplerDescriptor { - Texture now has `from_custom`. By @R-Cramer4 in [#8315](https://github.com/gfx-rs/wgpu/pull/8315). +### Added/New Features + +## General + +- Implement shader triangle barycentric coordinate builtins. By @atlv24 in [#8320](https://github.com/gfx-rs/wgpu/pull/8320). + ### Bug Fixes - Fixed bug where mapping sub-ranges of a buffer on web would fail with `OperationError: GPUBuffer.getMappedRange: GetMappedRange range extends beyond buffer's mapped range`. By @ryankaplan in [#8349](https://github.com/gfx-rs/wgpu/pull/8349) diff --git a/naga/src/back/glsl/features.rs b/naga/src/back/glsl/features.rs index a6dfe4e3100..4aef4ab6216 100644 --- a/naga/src/back/glsl/features.rs +++ b/naga/src/back/glsl/features.rs @@ -55,6 +55,8 @@ bitflags::bitflags! { const SUBGROUP_OPERATIONS = 1 << 24; /// Image atomics const TEXTURE_ATOMICS = 1 << 25; + /// Image atomics + const SHADER_BARYCENTRICS = 1 << 26; } } @@ -288,6 +290,14 @@ impl FeaturesManager { writeln!(out, "#extension GL_OES_shader_image_atomic : require")?; } + if self.0.contains(Features::SHADER_BARYCENTRICS) { + // https://github.com/KhronosGroup/GLSL/blob/main/extensions/ext/GLSL_EXT_fragment_shader_barycentric.txt + writeln!( + out, + "#extension GL_EXT_fragment_shader_barycentric : require" + )?; + } + Ok(()) } } @@ -603,6 +613,9 @@ impl Writer<'_, W> { crate::BuiltIn::InstanceIndex | crate::BuiltIn::DrawID => { self.features.request(Features::INSTANCE_INDEX) } + crate::BuiltIn::Barycentric => { + self.features.request(Features::SHADER_BARYCENTRICS) + } _ => {} }, Binding::Location { diff --git a/naga/src/back/glsl/mod.rs b/naga/src/back/glsl/mod.rs index 4c5a9d8cbcb..64c7714da58 100644 --- a/naga/src/back/glsl/mod.rs +++ b/naga/src/back/glsl/mod.rs @@ -5227,6 +5227,7 @@ const fn glsl_built_in(built_in: crate::BuiltIn, options: VaryingOptions) -> &'s Bi::PointCoord => "gl_PointCoord", Bi::FrontFacing => "gl_FrontFacing", Bi::PrimitiveIndex => "uint(gl_PrimitiveID)", + Bi::Barycentric => "gl_BaryCoordEXT", Bi::SampleIndex => "gl_SampleID", Bi::SampleMask => { if options.output { diff --git a/naga/src/back/hlsl/conv.rs b/naga/src/back/hlsl/conv.rs index ed40cbe5102..6cd8b398605 100644 --- a/naga/src/back/hlsl/conv.rs +++ b/naga/src/back/hlsl/conv.rs @@ -161,6 +161,7 @@ impl crate::BuiltIn { Self::FragDepth => "SV_Depth", Self::FrontFacing => "SV_IsFrontFace", Self::PrimitiveIndex => "SV_PrimitiveID", + Self::Barycentric => "SV_Barycentrics", Self::SampleIndex => "SV_SampleIndex", Self::SampleMask => "SV_Coverage", // compute diff --git a/naga/src/back/msl/mod.rs b/naga/src/back/msl/mod.rs index 44aedf686c4..fdbd434a57e 100644 --- a/naga/src/back/msl/mod.rs +++ b/naga/src/back/msl/mod.rs @@ -526,10 +526,15 @@ impl Options { return Err(Error::UnsupportedAttribute("instance_id".to_string())); } // macOS: Since Metal 2.2 - // iOS: Since Metal 2.3 (check depends on https://github.com/gfx-rs/naga/issues/2164) - crate::BuiltIn::PrimitiveIndex if self.lang_version < (2, 2) => { + // iOS: Since Metal 2.3 (check depends on https://github.com/gfx-rs/wgpu/issues/4414) + crate::BuiltIn::PrimitiveIndex if self.lang_version < (2, 3) => { return Err(Error::UnsupportedAttribute("primitive_id".to_string())); } + // macOS: Since Metal 2.2 + // iOS: Since Metal 2.3 (check depends on https://github.com/gfx-rs/wgpu/issues/4414) + crate::BuiltIn::Barycentric if self.lang_version < (2, 3) => { + return Err(Error::UnsupportedAttribute("barycentric_coord".to_string())); + } _ => {} } @@ -680,6 +685,7 @@ impl ResolvedBinding { Bi::PointCoord => "point_coord", Bi::FrontFacing => "front_facing", Bi::PrimitiveIndex => "primitive_id", + Bi::Barycentric => "barycentric_coord", Bi::SampleIndex => "sample_id", Bi::SampleMask => "sample_mask", // compute diff --git a/naga/src/back/spv/writer.rs b/naga/src/back/spv/writer.rs index c86a53c6ef8..6934ddb24b1 100644 --- a/naga/src/back/spv/writer.rs +++ b/naga/src/back/spv/writer.rs @@ -2089,6 +2089,14 @@ impl Writer { )?; BuiltIn::PrimitiveId } + Bi::Barycentric => { + self.require_any( + "`barycentric` built-in", + &[spirv::Capability::FragmentBarycentricKHR], + )?; + self.use_extension("SPV_KHR_fragment_shader_barycentric"); + BuiltIn::BaryCoordKHR + } Bi::SampleIndex => { self.require_any( "`sample_index` built-in", diff --git a/naga/src/common/wgsl/to_wgsl.rs b/naga/src/common/wgsl/to_wgsl.rs index 72be441288f..048585c054e 100644 --- a/naga/src/common/wgsl/to_wgsl.rs +++ b/naga/src/common/wgsl/to_wgsl.rs @@ -169,6 +169,7 @@ impl TryToWgsl for crate::BuiltIn { Bi::FragDepth => "frag_depth", Bi::FrontFacing => "front_facing", Bi::PrimitiveIndex => "primitive_index", + Bi::Barycentric => "barycentric", Bi::SampleIndex => "sample_index", Bi::SampleMask => "sample_mask", Bi::GlobalInvocationId => "global_invocation_id", diff --git a/naga/src/front/glsl/variables.rs b/naga/src/front/glsl/variables.rs index ef98143b769..003f3910a19 100644 --- a/naga/src/front/glsl/variables.rs +++ b/naga/src/front/glsl/variables.rs @@ -200,6 +200,7 @@ impl Frontend { "gl_BaseVertex" => BuiltIn::BaseVertex, "gl_BaseInstance" => BuiltIn::BaseInstance, "gl_PrimitiveID" => BuiltIn::PrimitiveIndex, + "gl_BaryCoordEXT" => BuiltIn::Barycentric, "gl_InstanceIndex" => BuiltIn::InstanceIndex, "gl_VertexIndex" => BuiltIn::VertexIndex, "gl_SampleID" => BuiltIn::SampleIndex, diff --git a/naga/src/front/spv/convert.rs b/naga/src/front/spv/convert.rs index 3e68c7bee21..7befa5f2938 100644 --- a/naga/src/front/spv/convert.rs +++ b/naga/src/front/spv/convert.rs @@ -147,6 +147,7 @@ pub(super) fn map_builtin(word: spirv::Word, invariant: bool) -> Result crate::BuiltIn::PointCoord, Some(Bi::FrontFacing) => crate::BuiltIn::FrontFacing, Some(Bi::PrimitiveId) => crate::BuiltIn::PrimitiveIndex, + Some(Bi::BaryCoordKHR) => crate::BuiltIn::Barycentric, Some(Bi::SampleId) => crate::BuiltIn::SampleIndex, Some(Bi::SampleMask) => crate::BuiltIn::SampleMask, // compute diff --git a/naga/src/front/spv/mod.rs b/naga/src/front/spv/mod.rs index 5e1b1146503..df6e09ce8ae 100644 --- a/naga/src/front/spv/mod.rs +++ b/naga/src/front/spv/mod.rs @@ -83,6 +83,7 @@ pub const SUPPORTED_CAPABILITIES: &[spirv::Capability] = &[ spirv::Capability::GroupNonUniformShuffleRelative, spirv::Capability::RuntimeDescriptorArray, spirv::Capability::StorageImageMultisample, + spirv::Capability::FragmentBarycentricKHR, // tricky ones spirv::Capability::UniformBufferArrayDynamicIndexing, spirv::Capability::StorageBufferArrayDynamicIndexing, @@ -6038,6 +6039,10 @@ impl> Frontend { size: crate::VectorSize::Tri, scalar: crate::Scalar::U32, }), + crate::BuiltIn::Barycentric => Some(crate::TypeInner::Vector { + size: crate::VectorSize::Tri, + scalar: crate::Scalar::F32, + }), _ => None, }; if let (Some(inner), Some(crate::ScalarKind::Sint)) = diff --git a/naga/src/front/wgsl/parse/conv.rs b/naga/src/front/wgsl/parse/conv.rs index 30d0eb2d598..de07ba2e391 100644 --- a/naga/src/front/wgsl/parse/conv.rs +++ b/naga/src/front/wgsl/parse/conv.rs @@ -36,6 +36,7 @@ pub fn map_built_in( "front_facing" => crate::BuiltIn::FrontFacing, "frag_depth" => crate::BuiltIn::FragDepth, "primitive_index" => crate::BuiltIn::PrimitiveIndex, + "barycentric" => crate::BuiltIn::Barycentric, "sample_index" => crate::BuiltIn::SampleIndex, "sample_mask" => crate::BuiltIn::SampleMask, // compute diff --git a/naga/src/ir/mod.rs b/naga/src/ir/mod.rs index 257445952b8..ecad643867c 100644 --- a/naga/src/ir/mod.rs +++ b/naga/src/ir/mod.rs @@ -387,6 +387,7 @@ pub enum BuiltIn { PointCoord, FrontFacing, PrimitiveIndex, + Barycentric, SampleIndex, SampleMask, // compute diff --git a/naga/src/valid/interface.rs b/naga/src/valid/interface.rs index 7c8cc903139..9d05447ad5d 100644 --- a/naga/src/valid/interface.rs +++ b/naga/src/valid/interface.rs @@ -180,6 +180,7 @@ impl VaryingContext<'_> { Bi::ClipDistance => Capabilities::CLIP_DISTANCE, Bi::CullDistance => Capabilities::CULL_DISTANCE, Bi::PrimitiveIndex => Capabilities::PRIMITIVE_INDEX, + Bi::Barycentric => Capabilities::SHADER_BARYCENTRICS, Bi::ViewIndex => Capabilities::MULTIVIEW, Bi::SampleIndex => Capabilities::MULTISAMPLED_SHADING, Bi::NumSubgroups @@ -267,6 +268,14 @@ impl VaryingContext<'_> { self.stage == St::Fragment && !self.output, *ty_inner == Ti::Scalar(crate::Scalar::U32), ), + Bi::Barycentric => ( + self.stage == St::Fragment && !self.output, + *ty_inner + == Ti::Vector { + size: Vs::Tri, + scalar: crate::Scalar::F32, + }, + ), Bi::SampleIndex => ( self.stage == St::Fragment && !self.output, *ty_inner == Ti::Scalar(crate::Scalar::U32), diff --git a/naga/src/valid/mod.rs b/naga/src/valid/mod.rs index 426b3d637d7..e68c0fa7a18 100644 --- a/naga/src/valid/mod.rs +++ b/naga/src/valid/mod.rs @@ -186,6 +186,8 @@ bitflags::bitflags! { /// Support for `quantizeToF16`, `pack2x16float`, and `unpack2x16float`, which store /// `f16`-precision values in `f32`s. const SHADER_FLOAT16_IN_FLOAT32 = 1 << 28; + /// Support for fragment shader barycentric coordinates. + const SHADER_BARYCENTRICS = 1 << 29; } } diff --git a/naga/tests/in/wgsl/barycentrics.toml b/naga/tests/in/wgsl/barycentrics.toml new file mode 100644 index 00000000000..d2cabe24309 --- /dev/null +++ b/naga/tests/in/wgsl/barycentrics.toml @@ -0,0 +1,10 @@ +god_mode = true + +[msl] +lang_version = [2, 3] + +[hlsl] +shader_model = "V6_1" + +[glsl] +version.Desktop = 450 diff --git a/naga/tests/in/wgsl/barycentrics.wgsl b/naga/tests/in/wgsl/barycentrics.wgsl new file mode 100644 index 00000000000..a87adaebefb --- /dev/null +++ b/naga/tests/in/wgsl/barycentrics.wgsl @@ -0,0 +1,4 @@ +@fragment +fn fs_main(@builtin(barycentric) bary: vec3) -> @location(0) vec4 { + return vec4(bary, 1.0); +} diff --git a/naga/tests/in/wgsl/extra.toml b/naga/tests/in/wgsl/extra.toml index e5e681a693a..8e940db7312 100644 --- a/naga/tests/in/wgsl/extra.toml +++ b/naga/tests/in/wgsl/extra.toml @@ -3,7 +3,7 @@ targets = "SPIRV | METAL | WGSL" [msl] fake_missing_bindings = false -lang_version = [2, 2] +lang_version = [2, 3] spirv_cross_compatibility = false zero_initialize_workgroup_memory = true diff --git a/naga/tests/naga/spirv_capabilities.rs b/naga/tests/naga/spirv_capabilities.rs index 3ac0efa2407..723d22d1e1e 100644 --- a/naga/tests/naga/spirv_capabilities.rs +++ b/naga/tests/naga/spirv_capabilities.rs @@ -148,6 +148,17 @@ fn sample_rate_shading() { ); } +#[test] +fn barycentrics() { + require( + &[Ca::FragmentBarycentricKHR], + r#" + @fragment + fn f(@builtin(barycentric) x: vec3) { } + "#, + ); +} + #[test] fn geometry() { require( diff --git a/naga/tests/out/glsl/wgsl-barycentrics.fs_main.Fragment.glsl b/naga/tests/out/glsl/wgsl-barycentrics.fs_main.Fragment.glsl new file mode 100644 index 00000000000..315eec530f8 --- /dev/null +++ b/naga/tests/out/glsl/wgsl-barycentrics.fs_main.Fragment.glsl @@ -0,0 +1,10 @@ +#version 450 core +#extension GL_EXT_fragment_shader_barycentric : require +layout(location = 0) out vec4 _fs2p_location0; + +void main() { + vec3 bary = gl_BaryCoordEXT; + _fs2p_location0 = vec4(bary, 1.0); + return; +} + diff --git a/naga/tests/out/hlsl/wgsl-barycentrics.hlsl b/naga/tests/out/hlsl/wgsl-barycentrics.hlsl new file mode 100644 index 00000000000..43b9cfdfe4e --- /dev/null +++ b/naga/tests/out/hlsl/wgsl-barycentrics.hlsl @@ -0,0 +1,9 @@ +struct FragmentInput_fs_main { + float3 bary_1 : SV_Barycentrics; +}; + +float4 fs_main(FragmentInput_fs_main fragmentinput_fs_main) : SV_Target0 +{ + float3 bary = fragmentinput_fs_main.bary_1; + return float4(bary, 1.0); +} diff --git a/naga/tests/out/hlsl/wgsl-barycentrics.ron b/naga/tests/out/hlsl/wgsl-barycentrics.ron new file mode 100644 index 00000000000..9ed6bdee58d --- /dev/null +++ b/naga/tests/out/hlsl/wgsl-barycentrics.ron @@ -0,0 +1,12 @@ +( + vertex:[ + ], + fragment:[ + ( + entry_point:"fs_main", + target_profile:"ps_6_1", + ), + ], + compute:[ + ], +) diff --git a/naga/tests/out/msl/wgsl-barycentrics.msl b/naga/tests/out/msl/wgsl-barycentrics.msl new file mode 100644 index 00000000000..1f539ba44db --- /dev/null +++ b/naga/tests/out/msl/wgsl-barycentrics.msl @@ -0,0 +1,17 @@ +// language: metal2.3 +#include +#include + +using metal::uint; + + +struct fs_mainInput { +}; +struct fs_mainOutput { + metal::float4 member [[color(0)]]; +}; +fragment fs_mainOutput fs_main( + metal::float3 bary [[barycentric_coord]] +) { + return fs_mainOutput { metal::float4(bary, 1.0) }; +} diff --git a/naga/tests/out/msl/wgsl-extra.msl b/naga/tests/out/msl/wgsl-extra.msl index be46447e286..43f83e648ae 100644 --- a/naga/tests/out/msl/wgsl-extra.msl +++ b/naga/tests/out/msl/wgsl-extra.msl @@ -1,4 +1,4 @@ -// language: metal2.2 +// language: metal2.3 #include #include diff --git a/naga/tests/out/spv/wgsl-barycentrics.spvasm b/naga/tests/out/spv/wgsl-barycentrics.spvasm new file mode 100644 index 00000000000..f497f000760 --- /dev/null +++ b/naga/tests/out/spv/wgsl-barycentrics.spvasm @@ -0,0 +1,32 @@ +; SPIR-V +; Version: 1.1 +; Generator: rspirv +; Bound: 17 +OpCapability Shader +OpCapability FragmentBarycentricKHR +OpExtension "SPV_KHR_fragment_shader_barycentric" +%1 = OpExtInstImport "GLSL.std.450" +OpMemoryModel Logical GLSL450 +OpEntryPoint Fragment %12 "fs_main" %7 %10 +OpExecutionMode %12 OriginUpperLeft +OpDecorate %7 BuiltIn BaryCoordKHR +OpDecorate %10 Location 0 +%2 = OpTypeVoid +%4 = OpTypeFloat 32 +%3 = OpTypeVector %4 3 +%5 = OpTypeVector %4 4 +%8 = OpTypePointer Input %3 +%7 = OpVariable %8 Input +%11 = OpTypePointer Output %5 +%10 = OpVariable %11 Output +%13 = OpTypeFunction %2 +%14 = OpConstant %4 1 +%12 = OpFunction %2 None %13 +%6 = OpLabel +%9 = OpLoad %3 %7 +OpBranch %15 +%15 = OpLabel +%16 = OpCompositeConstruct %5 %9 %14 +OpStore %10 %16 +OpReturn +OpFunctionEnd \ No newline at end of file diff --git a/naga/tests/out/wgsl/wgsl-barycentrics.wgsl b/naga/tests/out/wgsl/wgsl-barycentrics.wgsl new file mode 100644 index 00000000000..f34ccc23720 --- /dev/null +++ b/naga/tests/out/wgsl/wgsl-barycentrics.wgsl @@ -0,0 +1,4 @@ +@fragment +fn fs_main(@builtin(barycentric) bary: vec3) -> @location(0) vec4 { + return vec4(bary, 1f); +} diff --git a/tests/tests/wgpu-gpu/main.rs b/tests/tests/wgpu-gpu/main.rs index 82fe603c796..f6f967ffa1a 100644 --- a/tests/tests/wgpu-gpu/main.rs +++ b/tests/tests/wgpu-gpu/main.rs @@ -54,6 +54,7 @@ mod resource_error; mod samplers; mod scissor_tests; mod shader; +mod shader_barycentric; mod shader_primitive_index; mod shader_view_format; mod subgroup_operations; @@ -126,6 +127,7 @@ fn all_tests() -> Vec { samplers::all_tests(&mut tests); scissor_tests::all_tests(&mut tests); shader_primitive_index::all_tests(&mut tests); + shader_barycentric::all_tests(&mut tests); shader_view_format::all_tests(&mut tests); shader::all_tests(&mut tests); subgroup_operations::all_tests(&mut tests); diff --git a/tests/tests/wgpu-gpu/shader_barycentric/barycentric.wgsl b/tests/tests/wgpu-gpu/shader_barycentric/barycentric.wgsl new file mode 100644 index 00000000000..be3c6075bdf --- /dev/null +++ b/tests/tests/wgpu-gpu/shader_barycentric/barycentric.wgsl @@ -0,0 +1,9 @@ +@vertex +fn vs_main(@location(0) xy: vec2) -> @builtin(position) vec4 { + return vec4(xy, 0.0, 1.0); +} + +@fragment +fn fs_main(@builtin(barycentric) bary: vec3) -> @location(0) vec4 { + return vec4(bary * 1.1 - 0.05, 1.0); +} diff --git a/tests/tests/wgpu-gpu/shader_barycentric/mod.rs b/tests/tests/wgpu-gpu/shader_barycentric/mod.rs new file mode 100644 index 00000000000..b16d297e02c --- /dev/null +++ b/tests/tests/wgpu-gpu/shader_barycentric/mod.rs @@ -0,0 +1,164 @@ +use wgpu::util::DeviceExt; +use wgpu_test::{gpu_test, GpuTestConfiguration, TestParameters, TestingContext}; + +pub fn all_tests(vec: &mut Vec) { + vec.push(BARYCENTRIC); +} + +// +// This test renders one triangle to a 2x2 render target. The triangle +// covers the bottom-left, bottom-right, and the top-left pixel. +// XY layout of the render target, with the triangle: +// +// (-1,1) (0,1) (1,1) +// +------+------+ +// | | | +// | o | | +// | |\ | | +// | | \| | +// (-1,0) +---|--\------+ (1,0) +// | | |\ | +// | | | \ | +// | o--+--o | +// | | | +// +------+------+ +// (-1,-1) (0,-1) (1,-1) +// +// The fragment shader outputs color based on builtin(barycentric): +// +// return vec4(bary * 1.1 - 0.05, 1.0); +// + +#[gpu_test] +static BARYCENTRIC: GpuTestConfiguration = GpuTestConfiguration::new() + .parameters( + TestParameters::default() + .test_features_limits() + .features(wgpu::Features::SHADER_BARYCENTRICS), + ) + .run_async(barycentric); + +async fn barycentric(ctx: TestingContext) { + let shader = ctx + .device + .create_shader_module(wgpu::include_wgsl!("barycentric.wgsl")); + + let n = -0.505; + let p = 0.51; + let triangle_xy: [f32; 6] = [n, n, p, n, n, p]; + let vertex_buffer = ctx + .device + .create_buffer_init(&wgpu::util::BufferInitDescriptor { + label: None, + contents: bytemuck::cast_slice(&triangle_xy), + usage: wgpu::BufferUsages::VERTEX | wgpu::BufferUsages::COPY_DST, + }); + + let indices = [0u32, 1, 2]; + let index_buffer = ctx + .device + .create_buffer_init(&wgpu::util::BufferInitDescriptor { + label: None, + contents: bytemuck::cast_slice(&indices), + usage: wgpu::BufferUsages::INDEX | wgpu::BufferUsages::COPY_DST, + }); + + let pipeline = ctx + .device + .create_render_pipeline(&wgpu::RenderPipelineDescriptor { + label: None, + layout: None, + vertex: wgpu::VertexState { + module: &shader, + entry_point: Some("vs_main"), + compilation_options: Default::default(), + buffers: &[wgpu::VertexBufferLayout { + array_stride: 8, + step_mode: wgpu::VertexStepMode::Vertex, + attributes: &[wgpu::VertexAttribute { + format: wgpu::VertexFormat::Float32x2, + offset: 0, + shader_location: 0, + }], + }], + }, + primitive: wgpu::PrimitiveState::default(), + depth_stencil: None, + multisample: wgpu::MultisampleState::default(), + fragment: Some(wgpu::FragmentState { + module: &shader, + entry_point: Some("fs_main"), + compilation_options: Default::default(), + targets: &[Some(wgpu::ColorTargetState { + format: wgpu::TextureFormat::Rgba8Unorm, + blend: None, + write_mask: wgpu::ColorWrites::ALL, + })], + }), + multiview: None, + cache: None, + }); + + let width = 2; + let height = 2; + let texture_size = wgpu::Extent3d { + width, + height, + depth_or_array_layers: 1, + }; + let color_texture = ctx.device.create_texture(&wgpu::TextureDescriptor { + label: None, + size: texture_size, + mip_level_count: 1, + sample_count: 1, + dimension: wgpu::TextureDimension::D2, + format: wgpu::TextureFormat::Rgba8Unorm, + usage: wgpu::TextureUsages::RENDER_ATTACHMENT | wgpu::TextureUsages::COPY_SRC, + view_formats: &[], + }); + let color_view = color_texture.create_view(&wgpu::TextureViewDescriptor::default()); + + let readback_buffer = wgpu_test::image::ReadbackBuffers::new(&ctx.device, &color_texture); + + let mut encoder = ctx + .device + .create_command_encoder(&wgpu::CommandEncoderDescriptor::default()); + { + let mut rpass = encoder.begin_render_pass(&wgpu::RenderPassDescriptor { + label: None, + color_attachments: &[Some(wgpu::RenderPassColorAttachment { + ops: wgpu::Operations { + load: wgpu::LoadOp::Clear(wgpu::Color::WHITE), + store: wgpu::StoreOp::Store, + }, + resolve_target: None, + view: &color_view, + depth_slice: None, + })], + depth_stencil_attachment: None, + timestamp_writes: None, + occlusion_query_set: None, + }); + + rpass.set_pipeline(&pipeline); + rpass.set_index_buffer(index_buffer.slice(..), wgpu::IndexFormat::Uint32); + rpass.set_vertex_buffer(0, vertex_buffer.slice(..)); + rpass.draw(0..3, 0..1); + } + readback_buffer.copy_from(&ctx.device, &mut encoder, &color_texture); + ctx.queue.submit(Some(encoder.finish())); + + // + // +-----+-----+ + // |blue |white| + // +-----+-----+ + // | red |green| + // +-----+-----+ + // + let expected = [ + 0, 0, 255, 255, 255, 255, 255, 255, 255, 0, 0, 255, 0, 255, 0, 255, + ]; + readback_buffer + .assert_buffer_contents(&ctx, &expected) + .await; +} diff --git a/wgpu-core/src/device/mod.rs b/wgpu-core/src/device/mod.rs index 38e1b1d08f8..1ce57c2648e 100644 --- a/wgpu-core/src/device/mod.rs +++ b/wgpu-core/src/device/mod.rs @@ -510,6 +510,10 @@ pub fn create_validator( Caps::TEXTURE_EXTERNAL, features.intersects(wgt::Features::EXTERNAL_TEXTURE), ); + caps.set( + Caps::SHADER_BARYCENTRICS, + features.intersects(wgt::Features::SHADER_BARYCENTRICS), + ); naga::valid::Validator::new(flags, caps) } diff --git a/wgpu-hal/src/dx12/adapter.rs b/wgpu-hal/src/dx12/adapter.rs index e333d6a8824..f2c4c97978c 100644 --- a/wgpu-hal/src/dx12/adapter.rs +++ b/wgpu-hal/src/dx12/adapter.rs @@ -561,6 +561,23 @@ impl super::Adapter { wgt::Features::EXPERIMENTAL_MESH_SHADER, mesh_shader_supported, ); + let shader_barycentrics_supported = { + let mut features3 = Direct3D12::D3D12_FEATURE_DATA_D3D12_OPTIONS3::default(); + unsafe { + device.CheckFeatureSupport( + Direct3D12::D3D12_FEATURE_D3D12_OPTIONS3, + <*mut _>::cast(&mut features3), + size_of_val(&features3) as u32, + ) + } + .is_ok() + && features3.BarycentricsSupported.as_bool() + && shader_model >= naga::back::hlsl::ShaderModel::V6_1 + }; + features.set( + wgt::Features::SHADER_BARYCENTRICS, + shader_barycentrics_supported, + ); // TODO: Determine if IPresentationManager is supported let presentation_timer = auxil::dxgi::time::PresentationTimer::new_dxgi(); diff --git a/wgpu-hal/src/metal/adapter.rs b/wgpu-hal/src/metal/adapter.rs index 4e1be2509c1..20af4fdd0e9 100644 --- a/wgpu-hal/src/metal/adapter.rs +++ b/wgpu-hal/src/metal/adapter.rs @@ -902,6 +902,7 @@ impl super::PrivateCapabilities { && (device.supports_family(MTLGPUFamily::Apple7) || device.supports_family(MTLGPUFamily::Mac2)), supports_shared_event: version.at_least((10, 14), (12, 0), os_is_mac), + shader_barycentrics: device.supports_shader_barycentric_coordinates(), // https://developer.apple.com/metal/Metal-Feature-Set-Tables.pdf#page=3 supports_memoryless_storage: if family_check { device.supports_family(MTLGPUFamily::Apple2) @@ -1003,6 +1004,11 @@ impl super::PrivateCapabilities { features.set(F::RG11B10UFLOAT_RENDERABLE, self.format_rg11b10_all); + features.set( + F::SHADER_BARYCENTRICS, + self.shader_barycentrics && self.msl_version >= MTLLanguageVersion::V2_2, + ); + if self.supports_simd_scoped_operations { features.insert(F::SUBGROUP | F::SUBGROUP_BARRIER); } diff --git a/wgpu-hal/src/metal/mod.rs b/wgpu-hal/src/metal/mod.rs index 0dead249b8f..b89cd1d149f 100644 --- a/wgpu-hal/src/metal/mod.rs +++ b/wgpu-hal/src/metal/mod.rs @@ -301,6 +301,7 @@ struct PrivateCapabilities { int64_atomics: bool, float_atomics: bool, supports_shared_event: bool, + shader_barycentrics: bool, supports_memoryless_storage: bool, } diff --git a/wgpu-hal/src/vulkan/adapter.rs b/wgpu-hal/src/vulkan/adapter.rs index ba50eed76f0..93e77d74146 100644 --- a/wgpu-hal/src/vulkan/adapter.rs +++ b/wgpu-hal/src/vulkan/adapter.rs @@ -126,6 +126,9 @@ pub struct PhysicalDeviceFeatures { /// Features provided by `VK_KHR_shader_integer_dot_product`, promoted to Vulkan 1.3. shader_integer_dot_product: Option>, + + /// Features provided by `VK_KHR_fragment_shader_barycentric` + shader_barycentrics: Option>, } impl PhysicalDeviceFeatures { @@ -199,6 +202,9 @@ impl PhysicalDeviceFeatures { if let Some(ref mut feature) = self.shader_integer_dot_product { info = info.push_next(feature); } + if let Some(ref mut feature) = self.shader_barycentrics { + info = info.push_next(feature); + } info } @@ -533,6 +539,17 @@ impl PhysicalDeviceFeatures { } else { None }, + shader_barycentrics: if enabled_extensions + .contains(&khr::fragment_shader_barycentric::NAME) + { + let needed = requested_features.intersects(wgt::Features::SHADER_BARYCENTRICS); + Some( + vk::PhysicalDeviceFragmentShaderBarycentricFeaturesKHR::default() + .fragment_shader_barycentric(needed), + ) + } else { + None + }, } } @@ -667,6 +684,13 @@ impl PhysicalDeviceFeatures { ); } + if let Some(ref shader_barycentrics) = self.shader_barycentrics { + features.set( + F::SHADER_BARYCENTRICS, + shader_barycentrics.fragment_shader_barycentric != 0, + ); + } + //if caps.supports_extension(khr::sampler_mirror_clamp_to_edge::NAME) { //if caps.supports_extension(ext::sampler_filter_minmax::NAME) { features.set( @@ -1178,6 +1202,11 @@ impl PhysicalDeviceProperties { extensions.push(ext::mesh_shader::NAME); } + // Require `VK_KHR_fragment_shader_barycentric` if the associated feature was requested + if requested_features.intersects(wgt::Features::SHADER_BARYCENTRICS) { + extensions.push(khr::fragment_shader_barycentric::NAME); + } + extensions } @@ -1623,6 +1652,13 @@ impl super::InstanceShared { features2 = features2.push_next(next); } + if capabilities.supports_extension(khr::fragment_shader_barycentric::NAME) { + let next = features + .shader_barycentrics + .insert(vk::PhysicalDeviceFragmentShaderBarycentricFeaturesKHR::default()); + features2 = features2.push_next(next); + } + unsafe { get_device_properties.get_physical_device_features2(phd, &mut features2) }; features2.features } else { @@ -2112,6 +2148,10 @@ impl super::Adapter { capabilities.push(spv::Capability::ClipDistance); } + if features.intersects(wgt::Features::SHADER_BARYCENTRICS) { + capabilities.push(spv::Capability::FragmentBarycentricKHR); + } + let mut flags = spv::WriterFlags::empty(); flags.set( spv::WriterFlags::DEBUG, diff --git a/wgpu-types/src/features.rs b/wgpu-types/src/features.rs index 397de544d6d..c36a16c35ea 100644 --- a/wgpu-types/src/features.rs +++ b/wgpu-types/src/features.rs @@ -1231,6 +1231,16 @@ bitflags_array! { /// /// [`Device::create_shader_module_passthrough`]: https://docs.rs/wgpu/latest/wgpu/struct.Device.html#method.create_shader_module_passthrough const EXPERIMENTAL_PASSTHROUGH_SHADERS = 1 << 52; + + /// Enables shader barycentric coordinates. + /// + /// Supported platforms: + /// - Vulkan (with VK_KHR_fragment_shader_barycentric) + /// - DX12 (with SM 6.1+) + /// - Metal (with MSL 2.2+) + /// + /// This is a native only feature. + const SHADER_BARYCENTRICS = 1 << 53; } /// Features that are not guaranteed to be supported.