naga/back/msl/
mod.rs

1/*!
2Backend for [MSL][msl] (Metal Shading Language).
3
4This backend does not support the [`SHADER_INT64_ATOMIC_ALL_OPS`][all-atom]
5capability.
6
7## Binding model
8
9Metal's bindings are flat per resource. Since there isn't an obvious mapping
10from SPIR-V's descriptor sets, we require a separate mapping provided in the options.
11This mapping may have one or more resource end points for each descriptor set + index
12pair.
13
14## Entry points
15
16Even though MSL and our IR appear to be similar in that the entry points in both can
17accept arguments and return values, the restrictions are different.
18MSL allows the varyings to be either in separate arguments, or inside a single
19`[[stage_in]]` struct. We gather input varyings and form this artificial structure.
20We also add all the (non-Private) globals into the arguments.
21
22At the beginning of the entry point, we assign the local constants and re-compose
23the arguments as they are declared on IR side, so that the rest of the logic can
24pretend that MSL doesn't have all the restrictions it has.
25
26For the result type, if it's a structure, we re-compose it with a temporary value
27holding the result.
28
29[msl]: https://developer.apple.com/metal/Metal-Shading-Language-Specification.pdf
30[all-atom]: crate::valid::Capabilities::SHADER_INT64_ATOMIC_ALL_OPS
31
32## Pointer-typed bounds-checked expressions and OOB locals
33
34MSL (unlike HLSL and GLSL) has native support for pointer-typed function
35arguments. When the [`BoundsCheckPolicy`] is `ReadZeroSkipWrite` and an
36out-of-bounds index expression is used for such an argument, our strategy is to
37pass a pointer to a dummy variable. These dummy variables are called "OOB
38locals". We emit at most one OOB local per function for each type, since all
39expressions producing a result of that type can share the same OOB local. (Note
40that the OOB local mechanism is not actually implementing "skip write", nor even
41"read zero" in some cases of read-after-write, but doing so would require
42additional effort and the difference is unlikely to matter.)
43
44[`BoundsCheckPolicy`]: crate::proc::BoundsCheckPolicy
45
46## External textures
47
48Support for [`crate::ImageClass::External`] textures is implemented by lowering
49each external texture global variable to 3 `texture2d<float, sample>`s, and a
50constant buffer of type `NagaExternalTextureParams`. This provides up to 3
51planes of texture data (for example single planar RGBA, or separate Y, Cb, and
52Cr planes), and the parameters buffer containing information describing how to
53handle these correctly. The bind target to use for each of these globals is
54specified via the [`BindTarget::external_texture`] field of the relevant
55entries in [`EntryPointResources::resources`].
56
57External textures are supported by WGSL's `textureDimensions()`,
58`textureLoad()`, and `textureSampleBaseClampToEdge()` built-in functions. These
59are implemented using helper functions. See the following functions for how
60these are generated:
61 * `Writer::write_wrapped_image_query`
62 * `Writer::write_wrapped_image_load`
63 * `Writer::write_wrapped_image_sample`
64
65The lowered global variables for each external texture global are passed to the
66entry point as separate arguments (see "Entry points" above). However, they are
67then wrapped in a struct to allow them to be conveniently passed to user
68defined and helper functions. See `writer::EXTERNAL_TEXTURE_WRAPPER_STRUCT`.
69*/
70
71use alloc::{
72    format,
73    string::{String, ToString},
74    vec::Vec,
75};
76use core::fmt::{Error as FmtError, Write};
77
78use crate::{arena::Handle, ir, proc::index, valid::ModuleInfo};
79
80mod keywords;
81pub mod sampler;
82mod writer;
83
84pub use writer::Writer;
85
86pub type Slot = u8;
87pub type InlineSamplerIndex = u8;
88
89#[derive(Clone, Debug, PartialEq, Eq, Hash)]
90#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
91#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
92pub enum BindSamplerTarget {
93    Resource(Slot),
94    Inline(InlineSamplerIndex),
95}
96
97/// Binding information for a Naga [`External`] image global variable.
98///
99/// See the module documentation's section on external textures for details.
100///
101/// [`External`]: crate::ir::ImageClass::External
102#[derive(Clone, Debug, PartialEq, Eq, Hash)]
103#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
104#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
105pub struct BindExternalTextureTarget {
106    pub planes: [Slot; 3],
107    pub params: Slot,
108}
109
110#[derive(Clone, Debug, Default, PartialEq, Eq, Hash)]
111#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
112#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
113#[cfg_attr(any(feature = "serialize", feature = "deserialize"), serde(default))]
114pub struct BindTarget {
115    pub buffer: Option<Slot>,
116    pub texture: Option<Slot>,
117    pub sampler: Option<BindSamplerTarget>,
118    pub external_texture: Option<BindExternalTextureTarget>,
119    pub mutable: bool,
120}
121
122#[cfg(any(feature = "serialize", feature = "deserialize"))]
123#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
124#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
125struct BindingMapSerialization {
126    resource_binding: crate::ResourceBinding,
127    bind_target: BindTarget,
128}
129
130#[cfg(feature = "deserialize")]
131fn deserialize_binding_map<'de, D>(deserializer: D) -> Result<BindingMap, D::Error>
132where
133    D: serde::Deserializer<'de>,
134{
135    use serde::Deserialize;
136
137    let vec = Vec::<BindingMapSerialization>::deserialize(deserializer)?;
138    let mut map = BindingMap::default();
139    for item in vec {
140        map.insert(item.resource_binding, item.bind_target);
141    }
142    Ok(map)
143}
144
145// Using `BTreeMap` instead of `HashMap` so that we can hash itself.
146pub type BindingMap = alloc::collections::BTreeMap<crate::ResourceBinding, BindTarget>;
147
148#[derive(Clone, Debug, Default, Hash, Eq, PartialEq)]
149#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
150#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
151#[cfg_attr(any(feature = "serialize", feature = "deserialize"), serde(default))]
152pub struct EntryPointResources {
153    #[cfg_attr(
154        feature = "deserialize",
155        serde(deserialize_with = "deserialize_binding_map")
156    )]
157    pub resources: BindingMap,
158
159    pub push_constant_buffer: Option<Slot>,
160
161    /// The slot of a buffer that contains an array of `u32`,
162    /// one for the size of each bound buffer that contains a runtime array,
163    /// in order of [`crate::GlobalVariable`] declarations.
164    pub sizes_buffer: Option<Slot>,
165}
166
167pub type EntryPointResourceMap = alloc::collections::BTreeMap<String, EntryPointResources>;
168
169enum ResolvedBinding {
170    BuiltIn(crate::BuiltIn),
171    Attribute(u32),
172    Color {
173        location: u32,
174        blend_src: Option<u32>,
175    },
176    User {
177        prefix: &'static str,
178        index: u32,
179        interpolation: Option<ResolvedInterpolation>,
180    },
181    Resource(BindTarget),
182}
183
184#[derive(Copy, Clone)]
185enum ResolvedInterpolation {
186    CenterPerspective,
187    CenterNoPerspective,
188    CentroidPerspective,
189    CentroidNoPerspective,
190    SamplePerspective,
191    SampleNoPerspective,
192    Flat,
193}
194
195// Note: some of these should be removed in favor of proper IR validation.
196
197#[derive(Debug, thiserror::Error)]
198pub enum Error {
199    #[error(transparent)]
200    Format(#[from] FmtError),
201    #[error("bind target {0:?} is empty")]
202    UnimplementedBindTarget(BindTarget),
203    #[error("composing of {0:?} is not implemented yet")]
204    UnsupportedCompose(Handle<crate::Type>),
205    #[error("operation {0:?} is not implemented yet")]
206    UnsupportedBinaryOp(crate::BinaryOperator),
207    #[error("standard function '{0}' is not implemented yet")]
208    UnsupportedCall(String),
209    #[error("feature '{0}' is not implemented yet")]
210    FeatureNotImplemented(String),
211    #[error("internal naga error: module should not have validated: {0}")]
212    GenericValidation(String),
213    #[error("BuiltIn {0:?} is not supported")]
214    UnsupportedBuiltIn(crate::BuiltIn),
215    #[error("capability {0:?} is not supported")]
216    CapabilityNotSupported(crate::valid::Capabilities),
217    #[error("attribute '{0}' is not supported for target MSL version")]
218    UnsupportedAttribute(String),
219    #[error("function '{0}' is not supported for target MSL version")]
220    UnsupportedFunction(String),
221    #[error("can not use writeable storage buffers in fragment stage prior to MSL 1.2")]
222    UnsupportedWriteableStorageBuffer,
223    #[error("can not use writeable storage textures in {0:?} stage prior to MSL 1.2")]
224    UnsupportedWriteableStorageTexture(ir::ShaderStage),
225    #[error("can not use read-write storage textures prior to MSL 1.2")]
226    UnsupportedRWStorageTexture,
227    #[error("array of '{0}' is not supported for target MSL version")]
228    UnsupportedArrayOf(String),
229    #[error("array of type '{0:?}' is not supported")]
230    UnsupportedArrayOfType(Handle<crate::Type>),
231    #[error("ray tracing is not supported prior to MSL 2.3")]
232    UnsupportedRayTracing,
233    #[error("overrides should not be present at this stage")]
234    Override,
235    #[error("bitcasting to {0:?} is not supported")]
236    UnsupportedBitCast(crate::TypeInner),
237    #[error(transparent)]
238    ResolveArraySizeError(#[from] crate::proc::ResolveArraySizeError),
239    #[error("entry point with stage {0:?} and name '{1}' not found")]
240    EntryPointNotFound(ir::ShaderStage, String),
241}
242
243#[derive(Clone, Debug, PartialEq, thiserror::Error)]
244#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
245#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
246pub enum EntryPointError {
247    #[error("global '{0}' doesn't have a binding")]
248    MissingBinding(String),
249    #[error("mapping of {0:?} is missing")]
250    MissingBindTarget(crate::ResourceBinding),
251    #[error("mapping for push constants is missing")]
252    MissingPushConstants,
253    #[error("mapping for sizes buffer is missing")]
254    MissingSizesBuffer,
255}
256
257/// Points in the MSL code where we might emit a pipeline input or output.
258///
259/// Note that, even though vertex shaders' outputs are always fragment
260/// shaders' inputs, we still need to distinguish `VertexOutput` and
261/// `FragmentInput`, since there are certain differences in the way
262/// [`ResolvedBinding`s] are represented on either side.
263///
264/// [`ResolvedBinding`s]: ResolvedBinding
265#[derive(Clone, Copy, Debug)]
266enum LocationMode {
267    /// Input to the vertex shader.
268    VertexInput,
269
270    /// Output from the vertex shader.
271    VertexOutput,
272
273    /// Input to the fragment shader.
274    FragmentInput,
275
276    /// Output from the fragment shader.
277    FragmentOutput,
278
279    /// Compute shader input or output.
280    Uniform,
281}
282
283#[derive(Clone, Debug, Hash, PartialEq, Eq)]
284#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
285#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
286#[cfg_attr(feature = "deserialize", serde(default))]
287pub struct Options {
288    /// (Major, Minor) target version of the Metal Shading Language.
289    pub lang_version: (u8, u8),
290    /// Map of entry-point resources, indexed by entry point function name, to slots.
291    pub per_entry_point_map: EntryPointResourceMap,
292    /// Samplers to be inlined into the code.
293    pub inline_samplers: Vec<sampler::InlineSampler>,
294    /// Make it possible to link different stages via SPIRV-Cross.
295    pub spirv_cross_compatibility: bool,
296    /// Don't panic on missing bindings, instead generate invalid MSL.
297    pub fake_missing_bindings: bool,
298    /// Bounds checking policies.
299    pub bounds_check_policies: index::BoundsCheckPolicies,
300    /// Should workgroup variables be zero initialized (by polyfilling)?
301    pub zero_initialize_workgroup_memory: bool,
302    /// If set, loops will have code injected into them, forcing the compiler
303    /// to think the number of iterations is bounded.
304    pub force_loop_bounding: bool,
305}
306
307impl Default for Options {
308    fn default() -> Self {
309        Options {
310            lang_version: (1, 0),
311            per_entry_point_map: EntryPointResourceMap::default(),
312            inline_samplers: Vec::new(),
313            spirv_cross_compatibility: false,
314            fake_missing_bindings: true,
315            bounds_check_policies: index::BoundsCheckPolicies::default(),
316            zero_initialize_workgroup_memory: true,
317            force_loop_bounding: true,
318        }
319    }
320}
321
322/// Corresponds to [WebGPU `GPUVertexFormat`](
323/// https://gpuweb.github.io/gpuweb/#enumdef-gpuvertexformat).
324#[repr(u32)]
325#[derive(Copy, Clone, Debug, Hash, Eq, PartialEq)]
326#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
327#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
328pub enum VertexFormat {
329    /// One unsigned byte (u8). `u32` in shaders.
330    Uint8 = 0,
331    /// Two unsigned bytes (u8). `vec2<u32>` in shaders.
332    Uint8x2 = 1,
333    /// Four unsigned bytes (u8). `vec4<u32>` in shaders.
334    Uint8x4 = 2,
335    /// One signed byte (i8). `i32` in shaders.
336    Sint8 = 3,
337    /// Two signed bytes (i8). `vec2<i32>` in shaders.
338    Sint8x2 = 4,
339    /// Four signed bytes (i8). `vec4<i32>` in shaders.
340    Sint8x4 = 5,
341    /// One unsigned byte (u8). [0, 255] converted to float [0, 1] `f32` in shaders.
342    Unorm8 = 6,
343    /// Two unsigned bytes (u8). [0, 255] converted to float [0, 1] `vec2<f32>` in shaders.
344    Unorm8x2 = 7,
345    /// Four unsigned bytes (u8). [0, 255] converted to float [0, 1] `vec4<f32>` in shaders.
346    Unorm8x4 = 8,
347    /// One signed byte (i8). [-127, 127] converted to float [-1, 1] `f32` in shaders.
348    Snorm8 = 9,
349    /// Two signed bytes (i8). [-127, 127] converted to float [-1, 1] `vec2<f32>` in shaders.
350    Snorm8x2 = 10,
351    /// Four signed bytes (i8). [-127, 127] converted to float [-1, 1] `vec4<f32>` in shaders.
352    Snorm8x4 = 11,
353    /// One unsigned short (u16). `u32` in shaders.
354    Uint16 = 12,
355    /// Two unsigned shorts (u16). `vec2<u32>` in shaders.
356    Uint16x2 = 13,
357    /// Four unsigned shorts (u16). `vec4<u32>` in shaders.
358    Uint16x4 = 14,
359    /// One signed short (u16). `i32` in shaders.
360    Sint16 = 15,
361    /// Two signed shorts (i16). `vec2<i32>` in shaders.
362    Sint16x2 = 16,
363    /// Four signed shorts (i16). `vec4<i32>` in shaders.
364    Sint16x4 = 17,
365    /// One unsigned short (u16). [0, 65535] converted to float [0, 1] `f32` in shaders.
366    Unorm16 = 18,
367    /// Two unsigned shorts (u16). [0, 65535] converted to float [0, 1] `vec2<f32>` in shaders.
368    Unorm16x2 = 19,
369    /// Four unsigned shorts (u16). [0, 65535] converted to float [0, 1] `vec4<f32>` in shaders.
370    Unorm16x4 = 20,
371    /// One signed short (i16). [-32767, 32767] converted to float [-1, 1] `f32` in shaders.
372    Snorm16 = 21,
373    /// Two signed shorts (i16). [-32767, 32767] converted to float [-1, 1] `vec2<f32>` in shaders.
374    Snorm16x2 = 22,
375    /// Four signed shorts (i16). [-32767, 32767] converted to float [-1, 1] `vec4<f32>` in shaders.
376    Snorm16x4 = 23,
377    /// One half-precision float (no Rust equiv). `f32` in shaders.
378    Float16 = 24,
379    /// Two half-precision floats (no Rust equiv). `vec2<f32>` in shaders.
380    Float16x2 = 25,
381    /// Four half-precision floats (no Rust equiv). `vec4<f32>` in shaders.
382    Float16x4 = 26,
383    /// One single-precision float (f32). `f32` in shaders.
384    Float32 = 27,
385    /// Two single-precision floats (f32). `vec2<f32>` in shaders.
386    Float32x2 = 28,
387    /// Three single-precision floats (f32). `vec3<f32>` in shaders.
388    Float32x3 = 29,
389    /// Four single-precision floats (f32). `vec4<f32>` in shaders.
390    Float32x4 = 30,
391    /// One unsigned int (u32). `u32` in shaders.
392    Uint32 = 31,
393    /// Two unsigned ints (u32). `vec2<u32>` in shaders.
394    Uint32x2 = 32,
395    /// Three unsigned ints (u32). `vec3<u32>` in shaders.
396    Uint32x3 = 33,
397    /// Four unsigned ints (u32). `vec4<u32>` in shaders.
398    Uint32x4 = 34,
399    /// One signed int (i32). `i32` in shaders.
400    Sint32 = 35,
401    /// Two signed ints (i32). `vec2<i32>` in shaders.
402    Sint32x2 = 36,
403    /// Three signed ints (i32). `vec3<i32>` in shaders.
404    Sint32x3 = 37,
405    /// Four signed ints (i32). `vec4<i32>` in shaders.
406    Sint32x4 = 38,
407    /// Three unsigned 10-bit integers and one 2-bit integer, packed into a 32-bit integer (u32). [0, 1024] converted to float [0, 1] `vec4<f32>` in shaders.
408    #[cfg_attr(
409        any(feature = "serialize", feature = "deserialize"),
410        serde(rename = "unorm10-10-10-2")
411    )]
412    Unorm10_10_10_2 = 43,
413    /// Four unsigned 8-bit integers, packed into a 32-bit integer (u32). [0, 255] converted to float [0, 1] `vec4<f32>` in shaders.
414    #[cfg_attr(
415        any(feature = "serialize", feature = "deserialize"),
416        serde(rename = "unorm8x4-bgra")
417    )]
418    Unorm8x4Bgra = 44,
419}
420
421/// A mapping of vertex buffers and their attributes to shader
422/// locations.
423#[derive(Debug, Clone, PartialEq, Eq, Hash)]
424#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
425#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
426pub struct AttributeMapping {
427    /// Shader location associated with this attribute
428    pub shader_location: u32,
429    /// Offset in bytes from start of vertex buffer structure
430    pub offset: u32,
431    /// Format code to help us unpack the attribute into the type
432    /// used by the shader. Codes correspond to a 0-based index of
433    /// <https://gpuweb.github.io/gpuweb/#enumdef-gpuvertexformat>.
434    /// The conversion process is described by
435    /// <https://gpuweb.github.io/gpuweb/#vertex-processing>.
436    pub format: VertexFormat,
437}
438
439/// A description of a vertex buffer with all the information we
440/// need to address the attributes within it.
441#[derive(Debug, Default, Clone, PartialEq, Eq, Hash)]
442#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
443#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
444pub struct VertexBufferMapping {
445    /// Shader location associated with this buffer
446    pub id: u32,
447    /// Size of the structure in bytes
448    pub stride: u32,
449    /// True if the buffer is indexed by vertex, false if indexed
450    /// by instance.
451    pub indexed_by_vertex: bool,
452    /// Vec of the attributes within the structure
453    pub attributes: Vec<AttributeMapping>,
454}
455
456/// A subset of options that are meant to be changed per pipeline.
457#[derive(Debug, Default, Clone)]
458#[cfg_attr(feature = "serialize", derive(serde::Serialize))]
459#[cfg_attr(feature = "deserialize", derive(serde::Deserialize))]
460#[cfg_attr(feature = "deserialize", serde(default))]
461pub struct PipelineOptions {
462    /// The entry point to write.
463    ///
464    /// Entry points are identified by a shader stage specification,
465    /// and a name.
466    ///
467    /// If `None`, all entry points will be written. If `Some` and the entry
468    /// point is not found, an error will be thrown while writing.
469    pub entry_point: Option<(ir::ShaderStage, String)>,
470
471    /// Allow `BuiltIn::PointSize` and inject it if doesn't exist.
472    ///
473    /// Metal doesn't like this for non-point primitive topologies and requires it for
474    /// point primitive topologies.
475    ///
476    /// Enable this for vertex shaders with point primitive topologies.
477    pub allow_and_force_point_size: bool,
478
479    /// If set, when generating the Metal vertex shader, transform it
480    /// to receive the vertex buffers, lengths, and vertex id as args,
481    /// and bounds-check the vertex id and use the index into the
482    /// vertex buffers to access attributes, rather than using Metal's
483    /// [[stage-in]] assembled attribute data. This is true by default,
484    /// but remains configurable for use by tests via deserialization
485    /// of this struct. There is no user-facing way to set this value.
486    pub vertex_pulling_transform: bool,
487
488    /// vertex_buffer_mappings are used during shader translation to
489    /// support vertex pulling.
490    pub vertex_buffer_mappings: Vec<VertexBufferMapping>,
491}
492
493impl Options {
494    fn resolve_local_binding(
495        &self,
496        binding: &crate::Binding,
497        mode: LocationMode,
498    ) -> Result<ResolvedBinding, Error> {
499        match *binding {
500            crate::Binding::BuiltIn(mut built_in) => {
501                match built_in {
502                    crate::BuiltIn::Position { ref mut invariant } => {
503                        if *invariant && self.lang_version < (2, 1) {
504                            return Err(Error::UnsupportedAttribute("invariant".to_string()));
505                        }
506
507                        // The 'invariant' attribute may only appear on vertex
508                        // shader outputs, not fragment shader inputs.
509                        if !matches!(mode, LocationMode::VertexOutput) {
510                            *invariant = false;
511                        }
512                    }
513                    crate::BuiltIn::BaseInstance if self.lang_version < (1, 2) => {
514                        return Err(Error::UnsupportedAttribute("base_instance".to_string()));
515                    }
516                    crate::BuiltIn::InstanceIndex if self.lang_version < (1, 2) => {
517                        return Err(Error::UnsupportedAttribute("instance_id".to_string()));
518                    }
519                    // macOS: Since Metal 2.2
520                    // iOS: Since Metal 2.3 (check depends on https://github.com/gfx-rs/naga/issues/2164)
521                    crate::BuiltIn::PrimitiveIndex if self.lang_version < (2, 2) => {
522                        return Err(Error::UnsupportedAttribute("primitive_id".to_string()));
523                    }
524                    _ => {}
525                }
526
527                Ok(ResolvedBinding::BuiltIn(built_in))
528            }
529            crate::Binding::Location {
530                location,
531                interpolation,
532                sampling,
533                blend_src,
534            } => match mode {
535                LocationMode::VertexInput => Ok(ResolvedBinding::Attribute(location)),
536                LocationMode::FragmentOutput => {
537                    if blend_src.is_some() && self.lang_version < (1, 2) {
538                        return Err(Error::UnsupportedAttribute("blend_src".to_string()));
539                    }
540                    Ok(ResolvedBinding::Color {
541                        location,
542                        blend_src,
543                    })
544                }
545                LocationMode::VertexOutput | LocationMode::FragmentInput => {
546                    Ok(ResolvedBinding::User {
547                        prefix: if self.spirv_cross_compatibility {
548                            "locn"
549                        } else {
550                            "loc"
551                        },
552                        index: location,
553                        interpolation: {
554                            // unwrap: The verifier ensures that vertex shader outputs and fragment
555                            // shader inputs always have fully specified interpolation, and that
556                            // sampling is `None` only for Flat interpolation.
557                            let interpolation = interpolation.unwrap();
558                            let sampling = sampling.unwrap_or(crate::Sampling::Center);
559                            Some(ResolvedInterpolation::from_binding(interpolation, sampling))
560                        },
561                    })
562                }
563                LocationMode::Uniform => Err(Error::GenericValidation(format!(
564                    "Unexpected Binding::Location({location}) for the Uniform mode"
565                ))),
566            },
567        }
568    }
569
570    fn get_entry_point_resources(&self, ep: &crate::EntryPoint) -> Option<&EntryPointResources> {
571        self.per_entry_point_map.get(&ep.name)
572    }
573
574    fn get_resource_binding_target(
575        &self,
576        ep: &crate::EntryPoint,
577        res_binding: &crate::ResourceBinding,
578    ) -> Option<&BindTarget> {
579        self.get_entry_point_resources(ep)
580            .and_then(|res| res.resources.get(res_binding))
581    }
582
583    fn resolve_resource_binding(
584        &self,
585        ep: &crate::EntryPoint,
586        res_binding: &crate::ResourceBinding,
587    ) -> Result<ResolvedBinding, EntryPointError> {
588        let target = self.get_resource_binding_target(ep, res_binding);
589        match target {
590            Some(target) => Ok(ResolvedBinding::Resource(target.clone())),
591            None if self.fake_missing_bindings => Ok(ResolvedBinding::User {
592                prefix: "fake",
593                index: 0,
594                interpolation: None,
595            }),
596            None => Err(EntryPointError::MissingBindTarget(*res_binding)),
597        }
598    }
599
600    fn resolve_push_constants(
601        &self,
602        ep: &crate::EntryPoint,
603    ) -> Result<ResolvedBinding, EntryPointError> {
604        let slot = self
605            .get_entry_point_resources(ep)
606            .and_then(|res| res.push_constant_buffer);
607        match slot {
608            Some(slot) => Ok(ResolvedBinding::Resource(BindTarget {
609                buffer: Some(slot),
610                ..Default::default()
611            })),
612            None if self.fake_missing_bindings => Ok(ResolvedBinding::User {
613                prefix: "fake",
614                index: 0,
615                interpolation: None,
616            }),
617            None => Err(EntryPointError::MissingPushConstants),
618        }
619    }
620
621    fn resolve_sizes_buffer(
622        &self,
623        ep: &crate::EntryPoint,
624    ) -> Result<ResolvedBinding, EntryPointError> {
625        let slot = self
626            .get_entry_point_resources(ep)
627            .and_then(|res| res.sizes_buffer);
628        match slot {
629            Some(slot) => Ok(ResolvedBinding::Resource(BindTarget {
630                buffer: Some(slot),
631                ..Default::default()
632            })),
633            None if self.fake_missing_bindings => Ok(ResolvedBinding::User {
634                prefix: "fake",
635                index: 0,
636                interpolation: None,
637            }),
638            None => Err(EntryPointError::MissingSizesBuffer),
639        }
640    }
641}
642
643impl ResolvedBinding {
644    fn as_inline_sampler<'a>(&self, options: &'a Options) -> Option<&'a sampler::InlineSampler> {
645        match *self {
646            Self::Resource(BindTarget {
647                sampler: Some(BindSamplerTarget::Inline(index)),
648                ..
649            }) => Some(&options.inline_samplers[index as usize]),
650            _ => None,
651        }
652    }
653
654    fn try_fmt<W: Write>(&self, out: &mut W) -> Result<(), Error> {
655        write!(out, " [[")?;
656        match *self {
657            Self::BuiltIn(built_in) => {
658                use crate::BuiltIn as Bi;
659                let name = match built_in {
660                    Bi::Position { invariant: false } => "position",
661                    Bi::Position { invariant: true } => "position, invariant",
662                    // vertex
663                    Bi::BaseInstance => "base_instance",
664                    Bi::BaseVertex => "base_vertex",
665                    Bi::ClipDistance => "clip_distance",
666                    Bi::InstanceIndex => "instance_id",
667                    Bi::PointSize => "point_size",
668                    Bi::VertexIndex => "vertex_id",
669                    // fragment
670                    Bi::FragDepth => "depth(any)",
671                    Bi::PointCoord => "point_coord",
672                    Bi::FrontFacing => "front_facing",
673                    Bi::PrimitiveIndex => "primitive_id",
674                    Bi::SampleIndex => "sample_id",
675                    Bi::SampleMask => "sample_mask",
676                    // compute
677                    Bi::GlobalInvocationId => "thread_position_in_grid",
678                    Bi::LocalInvocationId => "thread_position_in_threadgroup",
679                    Bi::LocalInvocationIndex => "thread_index_in_threadgroup",
680                    Bi::WorkGroupId => "threadgroup_position_in_grid",
681                    Bi::WorkGroupSize => "dispatch_threads_per_threadgroup",
682                    Bi::NumWorkGroups => "threadgroups_per_grid",
683                    // subgroup
684                    Bi::NumSubgroups => "simdgroups_per_threadgroup",
685                    Bi::SubgroupId => "simdgroup_index_in_threadgroup",
686                    Bi::SubgroupSize => "threads_per_simdgroup",
687                    Bi::SubgroupInvocationId => "thread_index_in_simdgroup",
688                    Bi::CullDistance | Bi::ViewIndex | Bi::DrawID => {
689                        return Err(Error::UnsupportedBuiltIn(built_in))
690                    }
691                };
692                write!(out, "{name}")?;
693            }
694            Self::Attribute(index) => write!(out, "attribute({index})")?,
695            Self::Color {
696                location,
697                blend_src,
698            } => {
699                if let Some(blend_src) = blend_src {
700                    write!(out, "color({location}) index({blend_src})")?
701                } else {
702                    write!(out, "color({location})")?
703                }
704            }
705            Self::User {
706                prefix,
707                index,
708                interpolation,
709            } => {
710                write!(out, "user({prefix}{index})")?;
711                if let Some(interpolation) = interpolation {
712                    write!(out, ", ")?;
713                    interpolation.try_fmt(out)?;
714                }
715            }
716            Self::Resource(ref target) => {
717                if let Some(id) = target.buffer {
718                    write!(out, "buffer({id})")?;
719                } else if let Some(id) = target.texture {
720                    write!(out, "texture({id})")?;
721                } else if let Some(BindSamplerTarget::Resource(id)) = target.sampler {
722                    write!(out, "sampler({id})")?;
723                } else {
724                    return Err(Error::UnimplementedBindTarget(target.clone()));
725                }
726            }
727        }
728        write!(out, "]]")?;
729        Ok(())
730    }
731}
732
733impl ResolvedInterpolation {
734    const fn from_binding(interpolation: crate::Interpolation, sampling: crate::Sampling) -> Self {
735        use crate::Interpolation as I;
736        use crate::Sampling as S;
737
738        match (interpolation, sampling) {
739            (I::Perspective, S::Center) => Self::CenterPerspective,
740            (I::Perspective, S::Centroid) => Self::CentroidPerspective,
741            (I::Perspective, S::Sample) => Self::SamplePerspective,
742            (I::Linear, S::Center) => Self::CenterNoPerspective,
743            (I::Linear, S::Centroid) => Self::CentroidNoPerspective,
744            (I::Linear, S::Sample) => Self::SampleNoPerspective,
745            (I::Flat, _) => Self::Flat,
746            _ => unreachable!(),
747        }
748    }
749
750    fn try_fmt<W: Write>(self, out: &mut W) -> Result<(), Error> {
751        let identifier = match self {
752            Self::CenterPerspective => "center_perspective",
753            Self::CenterNoPerspective => "center_no_perspective",
754            Self::CentroidPerspective => "centroid_perspective",
755            Self::CentroidNoPerspective => "centroid_no_perspective",
756            Self::SamplePerspective => "sample_perspective",
757            Self::SampleNoPerspective => "sample_no_perspective",
758            Self::Flat => "flat",
759        };
760        out.write_str(identifier)?;
761        Ok(())
762    }
763}
764
765/// Information about a translated module that is required
766/// for the use of the result.
767pub struct TranslationInfo {
768    /// Mapping of the entry point names. Each item in the array
769    /// corresponds to an entry point index.
770    ///
771    ///Note: Some entry points may fail translation because of missing bindings.
772    pub entry_point_names: Vec<Result<String, EntryPointError>>,
773}
774
775pub fn write_string(
776    module: &crate::Module,
777    info: &ModuleInfo,
778    options: &Options,
779    pipeline_options: &PipelineOptions,
780) -> Result<(String, TranslationInfo), Error> {
781    let mut w = Writer::new(String::new());
782    let info = w.write(module, info, options, pipeline_options)?;
783    Ok((w.finish(), info))
784}
785
786#[test]
787fn test_error_size() {
788    assert_eq!(size_of::<Error>(), 40);
789}