diff --git a/src/backend/dx11/src/lib.rs b/src/backend/dx11/src/lib.rs index c24eef941d8..bbd0cec17d8 100644 --- a/src/backend/dx11/src/lib.rs +++ b/src/backend/dx11/src/lib.rs @@ -410,7 +410,14 @@ fn get_format_properties( props.buffer_features |= format::BufferFeature::STORAGE_TEXEL; } if can_image { + // Since read-only storage is exposed as SRV, we can guarantee read-only storage without checking D3D11_FORMAT_SUPPORT2_UAV_TYPED_LOAD first. props.optimal_tiling |= format::ImageFeature::STORAGE; + + if support_2.OutFormatSupport2 & d3d11::D3D11_FORMAT_SUPPORT2_UAV_TYPED_LOAD + != 0 + { + props.optimal_tiling |= format::ImageFeature::STORAGE_READ_WRITE; + } } } } diff --git a/src/backend/dx12/src/lib.rs b/src/backend/dx12/src/lib.rs index 82febee3cde..84bdad1eeb5 100644 --- a/src/backend/dx12/src/lib.rs +++ b/src/backend/dx12/src/lib.rs @@ -1456,7 +1456,12 @@ impl FormatProperties { props.buffer_features |= f::BufferFeature::STORAGE_TEXEL; } if can_image { + // Since read-only storage is exposed as SRV, we can guarantee read-only storage without checking D3D11_FORMAT_SUPPORT2_UAV_TYPED_LOAD first. props.optimal_tiling |= f::ImageFeature::STORAGE; + + if data.Support2 & d3d12::D3D12_FORMAT_SUPPORT2_UAV_TYPED_LOAD != 0 { + props.optimal_tiling |= f::ImageFeature::STORAGE_READ_WRITE; + } } } //TODO: blits, linear tiling diff --git a/src/backend/metal/src/conversions.rs b/src/backend/metal/src/conversions.rs index 02be6319310..7311b6315ac 100644 --- a/src/backend/metal/src/conversions.rs +++ b/src/backend/metal/src/conversions.rs @@ -179,6 +179,15 @@ impl PrivateCapabilities { let compressed_if = color_if | If::SAMPLED_LINEAR; let depth_if = color_if | If::DEPTH_STENCIL_ATTACHMENT; + // Affected formats documented at: + // https://developer.apple.com/documentation/metal/mtlreadwritetexturetier/mtlreadwritetexturetier1?language=objc + // https://developer.apple.com/documentation/metal/mtlreadwritetexturetier/mtlreadwritetexturetier2?language=objc + let (read_write_tier1_if, read_write_tier2_if) = match self.read_write_texture_tier { + MTLReadWriteTextureTier::TierNone => (If::empty(), If::empty()), + MTLReadWriteTextureTier::Tier1 => (If::STORAGE_READ_WRITE, If::empty()), + MTLReadWriteTextureTier::Tier2 => (If::STORAGE_READ_WRITE, If::STORAGE_READ_WRITE), + }; + match self.map_format(format) { Some(A8Unorm) => Properties { optimal_tiling: compressed_if, @@ -187,6 +196,7 @@ impl PrivateCapabilities { }, Some(R8Unorm) => Properties { optimal_tiling: color_if + | read_write_tier2_if | If::SAMPLED_LINEAR | If::STORAGE | If::COLOR_ATTACHMENT @@ -221,12 +231,12 @@ impl PrivateCapabilities { ..Properties::default() }, Some(R8Uint) => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier2_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, Some(R8Sint) => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier2_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, @@ -249,17 +259,18 @@ impl PrivateCapabilities { ..Properties::default() }, Some(R16Uint) => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier2_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, Some(R16Sint) => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier2_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, Some(R16Float) => Properties { optimal_tiling: color_if + | read_write_tier2_if | If::SAMPLED_LINEAR | If::STORAGE | If::COLOR_ATTACHMENT @@ -345,7 +356,7 @@ impl PrivateCapabilities { ..Properties::default() }, Some(R32Uint) if self.format_r32_all => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier1_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, @@ -355,7 +366,7 @@ impl PrivateCapabilities { ..Properties::default() }, Some(R32Sint) if self.format_r32_all => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier1_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, @@ -379,6 +390,7 @@ impl PrivateCapabilities { }, Some(R32Float) if self.format_r32float_all => Properties { optimal_tiling: color_if + | read_write_tier1_if | If::SAMPLED_LINEAR | If::STORAGE | If::COLOR_ATTACHMENT @@ -415,6 +427,7 @@ impl PrivateCapabilities { }, Some(RGBA8Unorm) => Properties { optimal_tiling: color_if + | read_write_tier2_if | If::SAMPLED_LINEAR | If::STORAGE | If::COLOR_ATTACHMENT @@ -449,12 +462,12 @@ impl PrivateCapabilities { ..Properties::default() }, Some(RGBA8Uint) => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier2_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, Some(RGBA8Sint) => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier2_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, @@ -611,17 +624,18 @@ impl PrivateCapabilities { ..Properties::default() }, Some(RGBA16Uint) => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier2_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, Some(RGBA16Sint) => Properties { - optimal_tiling: color_if | If::STORAGE | If::COLOR_ATTACHMENT, + optimal_tiling: color_if | read_write_tier2_if | If::STORAGE | If::COLOR_ATTACHMENT, buffer_features, ..Properties::default() }, Some(RGBA16Float) => Properties { optimal_tiling: color_if + | read_write_tier2_if | If::SAMPLED_LINEAR | If::STORAGE | If::COLOR_ATTACHMENT @@ -635,7 +649,7 @@ impl PrivateCapabilities { ..Properties::default() }, Some(RGBA32Uint) if self.format_rgba32int_color_write => Properties { - optimal_tiling: color_if | If::COLOR_ATTACHMENT | If::STORAGE, + optimal_tiling: color_if | read_write_tier2_if | If::COLOR_ATTACHMENT | If::STORAGE, buffer_features, ..Properties::default() }, @@ -645,12 +659,13 @@ impl PrivateCapabilities { ..Properties::default() }, Some(RGBA32Sint) if self.format_rgba32int_color_write => Properties { - optimal_tiling: color_if | If::COLOR_ATTACHMENT | If::STORAGE, + optimal_tiling: color_if | read_write_tier2_if | If::COLOR_ATTACHMENT | If::STORAGE, buffer_features, ..Properties::default() }, Some(RGBA32Float) if self.format_rgba32float_all => Properties { optimal_tiling: color_if + | read_write_tier2_if | If::SAMPLED_LINEAR | If::STORAGE | If::COLOR_ATTACHMENT @@ -664,7 +679,7 @@ impl PrivateCapabilities { ..Properties::default() }, Some(RGBA32Float) if self.format_rgba32float_color_write => Properties { - optimal_tiling: color_if | If::COLOR_ATTACHMENT | If::STORAGE, + optimal_tiling: color_if | read_write_tier2_if | If::COLOR_ATTACHMENT | If::STORAGE, buffer_features, ..Properties::default() }, diff --git a/src/backend/metal/src/lib.rs b/src/backend/metal/src/lib.rs index 987ca896651..f2d28c4c79f 100644 --- a/src/backend/metal/src/lib.rs +++ b/src/backend/metal/src/lib.rs @@ -680,6 +680,7 @@ struct PrivateCapabilities { os_version: (u32, u32), msl_version: metal::MTLLanguageVersion, exposed_queues: usize, + read_write_texture_tier: metal::MTLReadWriteTextureTier, // if TRUE, we'll report `NON_FILL_POLYGON_MODE` feature without the points support expose_line_mode: bool, resource_heaps: bool, @@ -823,6 +824,7 @@ impl PrivateCapabilities { MTLLanguageVersion::V1_0 }, exposed_queues: 1, + read_write_texture_tier: device.read_write_texture_support(), expose_line_mode: true, resource_heaps: Self::supports_any(&device, RESOURCE_HEAP_SUPPORT), argument_buffers: experiments.argument_buffers diff --git a/src/backend/vulkan/src/conv.rs b/src/backend/vulkan/src/conv.rs index 342b18738ac..3f808a3ecf5 100644 --- a/src/backend/vulkan/src/conv.rs +++ b/src/backend/vulkan/src/conv.rs @@ -378,7 +378,40 @@ pub fn map_query_result_flags(flags: query::ResultFlags) -> vk::QueryResultFlags } pub fn map_image_features(features: vk::FormatFeatureFlags) -> format::ImageFeature { - format::ImageFeature::from_bits_truncate(features.as_raw()) + let mut mapped_flags = format::ImageFeature::empty(); + if features.contains(vk::FormatFeatureFlags::SAMPLED_IMAGE) { + mapped_flags |= format::ImageFeature::SAMPLED; + } + if features.contains(vk::FormatFeatureFlags::SAMPLED_IMAGE_FILTER_LINEAR) { + mapped_flags |= format::ImageFeature::SAMPLED_LINEAR; + } + + if features.contains(vk::FormatFeatureFlags::STORAGE_IMAGE) { + mapped_flags |= format::ImageFeature::STORAGE; + mapped_flags |= format::ImageFeature::STORAGE_READ_WRITE; + } + if features.contains(vk::FormatFeatureFlags::STORAGE_IMAGE_ATOMIC) { + mapped_flags |= format::ImageFeature::STORAGE_ATOMIC; + } + + if features.contains(vk::FormatFeatureFlags::COLOR_ATTACHMENT) { + mapped_flags |= format::ImageFeature::COLOR_ATTACHMENT; + } + if features.contains(vk::FormatFeatureFlags::COLOR_ATTACHMENT_BLEND) { + mapped_flags |= format::ImageFeature::COLOR_ATTACHMENT_BLEND; + } + if features.contains(vk::FormatFeatureFlags::DEPTH_STENCIL_ATTACHMENT) { + mapped_flags |= format::ImageFeature::DEPTH_STENCIL_ATTACHMENT; + } + + if features.contains(vk::FormatFeatureFlags::BLIT_SRC) { + mapped_flags |= format::ImageFeature::BLIT_SRC; + } + if features.contains(vk::FormatFeatureFlags::BLIT_DST) { + mapped_flags |= format::ImageFeature::BLIT_DST; + } + + mapped_flags } pub fn map_buffer_features(features: vk::FormatFeatureFlags) -> format::BufferFeature { diff --git a/src/hal/src/format.rs b/src/hal/src/format.rs index 2415b57cf23..7aff481b3e0 100644 --- a/src/hal/src/format.rs +++ b/src/hal/src/format.rs @@ -148,24 +148,28 @@ bitflags!( pub struct ImageFeature: u32 { /// Image view can be sampled. const SAMPLED = 0x1; - /// Image view can be used as storage image. - const STORAGE = 0x2; - /// Image view can be used as storage image (with atomics). - const STORAGE_ATOMIC = 0x4; + /// Image can be sampled with a (mipmap) linear sampler or as blit source with linear sampling. + /// (implies SAMPLED and BLIT_SRC support) + const SAMPLED_LINEAR = 0x2; + + /// Image view can be used as storage image with exclusive read & write access. + const STORAGE = 0x10; + /// Image view can be used as storage image with simultaneous read/write access. + const STORAGE_READ_WRITE = 0x20; + /// Image view can be used as storage image with atomics. + const STORAGE_ATOMIC = 0x40; + /// Image view can be used as color and input attachment. - const COLOR_ATTACHMENT = 0x80; + const COLOR_ATTACHMENT = 0x100; /// Image view can be used as color (with blending) and input attachment. - const COLOR_ATTACHMENT_BLEND = 0x100; + const COLOR_ATTACHMENT_BLEND = 0x200; /// Image view can be used as depth-stencil and input attachment. - const DEPTH_STENCIL_ATTACHMENT = 0x200; + const DEPTH_STENCIL_ATTACHMENT = 0x400; + /// Image can be used as source for blit commands. - const BLIT_SRC = 0x400; + const BLIT_SRC = 0x1000; /// Image can be used as destination for blit commands. - const BLIT_DST = 0x800; - /// Image can be sampled with a (mipmap) linear sampler or as blit source - /// with linear sampling. - /// Requires `SAMPLED` or `BLIT_SRC` flag. - const SAMPLED_LINEAR = 0x1000; + const BLIT_DST = 0x2000; } );