naga/valid/
interface.rs

1use alloc::vec::Vec;
2
3use bit_set::BitSet;
4
5use super::{
6    analyzer::{FunctionInfo, GlobalUse},
7    Capabilities, Disalignment, FunctionError, ImmediateError, ModuleInfo,
8};
9use crate::arena::{Handle, UniqueArena};
10use crate::span::{AddSpan as _, MapErrWithSpan as _, SpanProvider as _, WithSpan};
11
12const MAX_WORKGROUP_SIZE: u32 = 0x4000;
13
14#[derive(Clone, Debug, thiserror::Error)]
15#[cfg_attr(test, derive(PartialEq))]
16pub enum GlobalVariableError {
17    #[error("Usage isn't compatible with address space {0:?}")]
18    InvalidUsage(crate::AddressSpace),
19    #[error("Type isn't compatible with address space {0:?}")]
20    InvalidType(crate::AddressSpace),
21    #[error("Type {0:?} isn't compatible with binding arrays")]
22    InvalidBindingArray(Handle<crate::Type>),
23    #[error("Type flags {seen:?} do not meet the required {required:?}")]
24    MissingTypeFlags {
25        required: super::TypeFlags,
26        seen: super::TypeFlags,
27    },
28    #[error("Capability {0:?} is not supported")]
29    UnsupportedCapability(Capabilities),
30    #[error("Binding decoration is missing or not applicable")]
31    InvalidBinding,
32    #[error("Alignment requirements for address space {0:?} are not met by {1:?}")]
33    Alignment(
34        crate::AddressSpace,
35        Handle<crate::Type>,
36        #[source] Disalignment,
37    ),
38    #[error("Initializer must be an override-expression")]
39    InitializerExprType,
40    #[error("Initializer doesn't match the variable type")]
41    InitializerType,
42    #[error("Initializer can't be used with address space {0:?}")]
43    InitializerNotAllowed(crate::AddressSpace),
44    #[error("Storage address space doesn't support write-only access")]
45    StorageAddressSpaceWriteOnlyNotSupported,
46    #[error("Type is not valid for use as a immediate data")]
47    InvalidImmediateType(#[source] ImmediateError),
48    #[error("Task payload must not be zero-sized")]
49    ZeroSizedTaskPayload,
50}
51
52#[derive(Clone, Debug, thiserror::Error)]
53#[cfg_attr(test, derive(PartialEq))]
54pub enum VaryingError {
55    #[error("The type {0:?} does not match the varying")]
56    InvalidType(Handle<crate::Type>),
57    #[error("The type {0:?} cannot be used for user-defined entry point inputs or outputs")]
58    NotIOShareableType(Handle<crate::Type>),
59    #[error("Interpolation is not valid")]
60    InvalidInterpolation,
61    #[error("Interpolation {0:?} is only valid for stage {1:?}")]
62    InvalidInterpolationInStage(crate::Interpolation, crate::ShaderStage),
63    #[error("Cannot combine {interpolation:?} interpolation with the {sampling:?} sample type")]
64    InvalidInterpolationSamplingCombination {
65        interpolation: crate::Interpolation,
66        sampling: crate::Sampling,
67    },
68    #[error("Interpolation must be specified on vertex shader outputs and fragment shader inputs")]
69    MissingInterpolation,
70    #[error("Built-in {0:?} is not available at this stage")]
71    InvalidBuiltInStage(crate::BuiltIn),
72    #[error("Built-in type for {0:?} is invalid. Found {1:?}")]
73    InvalidBuiltInType(crate::BuiltIn, crate::TypeInner),
74    #[error("Entry point arguments and return values must all have bindings")]
75    MissingBinding,
76    #[error("Struct member {0} is missing a binding")]
77    MemberMissingBinding(u32),
78    #[error("Multiple bindings at location {location} are present")]
79    BindingCollision { location: u32 },
80    #[error("Multiple bindings use the same `blend_src` {blend_src}")]
81    BindingCollisionBlendSrc { blend_src: u32 },
82    #[error("Built-in {0:?} is present more than once")]
83    DuplicateBuiltIn(crate::BuiltIn),
84    #[error("Capability {0:?} is not supported")]
85    UnsupportedCapability(Capabilities),
86    #[error("The attribute {0:?} is only valid as an output for stage {1:?}")]
87    InvalidInputAttributeInStage(&'static str, crate::ShaderStage),
88    #[error("The attribute {0:?} is not valid for stage {1:?}")]
89    InvalidAttributeInStage(&'static str, crate::ShaderStage),
90    #[error("The `blend_src` attribute can only be used on location 0, only indices 0 and 1 are valid. Location was {location}, index was {blend_src}.")]
91    InvalidBlendSrcIndex { location: u32, blend_src: u32 },
92    #[error("If `blend_src` is used, there must be exactly two outputs both with location 0, one with `blend_src(0)` and the other with `blend_src(1)`.")]
93    IncompleteBlendSrcUsage,
94    #[error("If `blend_src` is used, both outputs must have the same type. `blend_src(0)` has type {blend_src_0_type:?} and `blend_src(1)` has type {blend_src_1_type:?}.")]
95    BlendSrcOutputTypeMismatch {
96        blend_src_0_type: Handle<crate::Type>,
97        blend_src_1_type: Handle<crate::Type>,
98    },
99    #[error("Workgroup size is multi dimensional, `@builtin(subgroup_id)` and `@builtin(subgroup_invocation_id)` are not supported.")]
100    InvalidMultiDimensionalSubgroupBuiltIn,
101    #[error("The `@per_primitive` attribute can only be used in fragment shader inputs or mesh shader primitive outputs")]
102    InvalidPerPrimitive,
103    #[error("Non-builtin members of a mesh primitive output struct must be decorated with `@per_primitive`")]
104    MissingPerPrimitive,
105    #[error("Per vertex fragment inputs must be an array of length 3.")]
106    PerVertexNotArrayOfThree,
107}
108
109#[derive(Clone, Debug, thiserror::Error)]
110#[cfg_attr(test, derive(PartialEq))]
111pub enum EntryPointError {
112    #[error("Multiple conflicting entry points")]
113    Conflict,
114    #[error("Vertex shaders must return a `@builtin(position)` output value")]
115    MissingVertexOutputPosition,
116    #[error("Early depth test is not applicable")]
117    UnexpectedEarlyDepthTest,
118    #[error("Workgroup size is not applicable")]
119    UnexpectedWorkgroupSize,
120    #[error("Workgroup size is out of range")]
121    OutOfRangeWorkgroupSize,
122    #[error("Uses operations forbidden at this stage")]
123    ForbiddenStageOperations,
124    #[error("Global variable {0:?} is used incorrectly as {1:?}")]
125    InvalidGlobalUsage(Handle<crate::GlobalVariable>, GlobalUse),
126    #[error("More than 1 immediate data variable is used")]
127    MoreThanOneImmediateUsed,
128    #[error("Bindings for {0:?} conflict with other resource")]
129    BindingCollision(Handle<crate::GlobalVariable>),
130    #[error("Argument {0} varying error")]
131    Argument(u32, #[source] VaryingError),
132    #[error(transparent)]
133    Result(#[from] VaryingError),
134    #[error("Location {location} interpolation of an integer has to be flat")]
135    InvalidIntegerInterpolation { location: u32 },
136    #[error(transparent)]
137    Function(#[from] FunctionError),
138    #[error("Capability {0:?} is not supported")]
139    UnsupportedCapability(Capabilities),
140
141    #[error("mesh shader entry point missing mesh shader attributes")]
142    ExpectedMeshShaderAttributes,
143    #[error("Non mesh shader entry point cannot have mesh shader attributes")]
144    UnexpectedMeshShaderAttributes,
145    #[error("Non mesh/task shader entry point cannot have task payload attribute")]
146    UnexpectedTaskPayload,
147    #[error("Task payload must be declared with `var<task_payload>`")]
148    TaskPayloadWrongAddressSpace,
149    #[error("For a task payload to be used, it must be declared with @payload")]
150    WrongTaskPayloadUsed,
151    #[error("Task shader entry point must return @builtin(mesh_task_size) vec3<u32>")]
152    WrongTaskShaderEntryResult,
153    #[error("Task shaders must declare a task payload output")]
154    ExpectedTaskPayload,
155    #[error(
156        "Mesh shader output variable must be a struct with fields that are all allowed builtins"
157    )]
158    BadMeshOutputVariableType,
159    #[error("Mesh shader output variable fields must have types that are in accordance with the mesh shader spec")]
160    BadMeshOutputVariableField,
161    #[error("Mesh shader entry point cannot have a return type")]
162    UnexpectedMeshShaderEntryResult,
163    #[error(
164        "Mesh output type must be a user-defined struct with fields in alignment with the mesh shader spec"
165    )]
166    InvalidMeshOutputType,
167    #[error("Mesh primitive outputs must have exactly one of `@builtin(triangle_indices)`, `@builtin(line_indices)`, or `@builtin(point_index)`")]
168    InvalidMeshPrimitiveOutputType,
169    #[error("Mesh output global variable must live in the workgroup address space")]
170    WrongMeshOutputAddressSpace,
171    #[error("Task payload must be at least 4 bytes, but is {0} bytes")]
172    TaskPayloadTooSmall(u32),
173}
174
175fn storage_usage(access: crate::StorageAccess) -> GlobalUse {
176    let mut storage_usage = GlobalUse::QUERY;
177    if access.contains(crate::StorageAccess::LOAD) {
178        storage_usage |= GlobalUse::READ;
179    }
180    if access.contains(crate::StorageAccess::STORE) {
181        storage_usage |= GlobalUse::WRITE;
182    }
183    if access.contains(crate::StorageAccess::ATOMIC) {
184        storage_usage |= GlobalUse::ATOMIC;
185    }
186    storage_usage
187}
188
189#[derive(Clone, Copy, Debug, PartialEq, Eq)]
190enum MeshOutputType {
191    None,
192    VertexOutput,
193    PrimitiveOutput,
194}
195
196struct VaryingContext<'a> {
197    stage: crate::ShaderStage,
198    output: bool,
199    types: &'a UniqueArena<crate::Type>,
200    type_info: &'a Vec<super::r#type::TypeInfo>,
201    location_mask: &'a mut BitSet,
202    blend_src_mask: &'a mut BitSet,
203    built_ins: &'a mut crate::FastHashSet<crate::BuiltIn>,
204    capabilities: Capabilities,
205    flags: super::ValidationFlags,
206    mesh_output_type: MeshOutputType,
207    has_task_payload: bool,
208}
209
210impl VaryingContext<'_> {
211    fn validate_impl(
212        &mut self,
213        ep: &crate::EntryPoint,
214        ty: Handle<crate::Type>,
215        binding: &crate::Binding,
216    ) -> Result<(), VaryingError> {
217        use crate::{BuiltIn as Bi, ShaderStage as St, TypeInner as Ti, VectorSize as Vs};
218
219        let ty_inner = &self.types[ty].inner;
220        match *binding {
221            crate::Binding::BuiltIn(built_in) => {
222                // Ignore the `invariant` field for the sake of duplicate checks,
223                // but use the original in error messages.
224                let canonical = match built_in {
225                    crate::BuiltIn::Position { .. } => {
226                        crate::BuiltIn::Position { invariant: false }
227                    }
228                    crate::BuiltIn::Barycentric { .. } => {
229                        crate::BuiltIn::Barycentric { perspective: false }
230                    }
231                    x => x,
232                };
233
234                if self.built_ins.contains(&canonical) {
235                    return Err(VaryingError::DuplicateBuiltIn(built_in));
236                }
237                self.built_ins.insert(canonical);
238
239                let required = match built_in {
240                    Bi::ClipDistance => Capabilities::CLIP_DISTANCE,
241                    Bi::CullDistance => Capabilities::CULL_DISTANCE,
242                    Bi::PrimitiveIndex => Capabilities::PRIMITIVE_INDEX,
243                    Bi::Barycentric { .. } => Capabilities::SHADER_BARYCENTRICS,
244                    Bi::ViewIndex => Capabilities::MULTIVIEW,
245                    Bi::SampleIndex => Capabilities::MULTISAMPLED_SHADING,
246                    Bi::NumSubgroups
247                    | Bi::SubgroupId
248                    | Bi::SubgroupSize
249                    | Bi::SubgroupInvocationId => Capabilities::SUBGROUP,
250                    _ => Capabilities::empty(),
251                };
252                if !self.capabilities.contains(required) {
253                    return Err(VaryingError::UnsupportedCapability(required));
254                }
255
256                if matches!(
257                    built_in,
258                    crate::BuiltIn::SubgroupId | crate::BuiltIn::SubgroupInvocationId
259                ) && ep.workgroup_size[1..].iter().any(|&s| s > 1)
260                {
261                    return Err(VaryingError::InvalidMultiDimensionalSubgroupBuiltIn);
262                }
263
264                let (visible, type_good) = match built_in {
265                    Bi::BaseInstance | Bi::BaseVertex | Bi::InstanceIndex | Bi::VertexIndex => (
266                        self.stage == St::Vertex && !self.output,
267                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
268                    ),
269                    Bi::DrawID => (
270                        // Always allowed in task/vertex stage. Allowed in mesh stage if there is no task stage in the pipeline.
271                        (self.stage == St::Vertex
272                            || self.stage == St::Task
273                            || (self.stage == St::Mesh && !self.has_task_payload))
274                            && !self.output,
275                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
276                    ),
277                    Bi::ClipDistance | Bi::CullDistance => (
278                        (self.stage == St::Vertex || self.stage == St::Mesh) && self.output,
279                        match *ty_inner {
280                            Ti::Array { base, size, .. } => {
281                                self.types[base].inner == Ti::Scalar(crate::Scalar::F32)
282                                    && match size {
283                                        crate::ArraySize::Constant(non_zero) => non_zero.get() <= 8,
284                                        _ => false,
285                                    }
286                            }
287                            _ => false,
288                        },
289                    ),
290                    Bi::PointSize => (
291                        (self.stage == St::Vertex || self.stage == St::Mesh) && self.output,
292                        *ty_inner == Ti::Scalar(crate::Scalar::F32),
293                    ),
294                    Bi::PointCoord => (
295                        self.stage == St::Fragment && !self.output,
296                        *ty_inner
297                            == Ti::Vector {
298                                size: Vs::Bi,
299                                scalar: crate::Scalar::F32,
300                            },
301                    ),
302                    Bi::Position { .. } => (
303                        match self.stage {
304                            St::Vertex | St::Mesh => self.output,
305                            St::Fragment => !self.output,
306                            St::Compute | St::Task => false,
307                        },
308                        *ty_inner
309                            == Ti::Vector {
310                                size: Vs::Quad,
311                                scalar: crate::Scalar::F32,
312                            },
313                    ),
314                    Bi::ViewIndex => (
315                        match self.stage {
316                            St::Vertex | St::Fragment | St::Task | St::Mesh => !self.output,
317                            St::Compute => false,
318                        },
319                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
320                    ),
321                    Bi::FragDepth => (
322                        self.stage == St::Fragment && self.output,
323                        *ty_inner == Ti::Scalar(crate::Scalar::F32),
324                    ),
325                    Bi::FrontFacing => (
326                        self.stage == St::Fragment && !self.output,
327                        *ty_inner == Ti::Scalar(crate::Scalar::BOOL),
328                    ),
329                    Bi::PrimitiveIndex => (
330                        (self.stage == St::Fragment && !self.output)
331                            || (self.stage == St::Mesh
332                                && self.output
333                                && self.mesh_output_type == MeshOutputType::PrimitiveOutput),
334                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
335                    ),
336                    Bi::Barycentric { .. } => (
337                        self.stage == St::Fragment && !self.output,
338                        *ty_inner
339                            == Ti::Vector {
340                                size: Vs::Tri,
341                                scalar: crate::Scalar::F32,
342                            },
343                    ),
344                    Bi::SampleIndex => (
345                        self.stage == St::Fragment && !self.output,
346                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
347                    ),
348                    Bi::SampleMask => (
349                        self.stage == St::Fragment,
350                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
351                    ),
352                    Bi::LocalInvocationIndex => (
353                        self.stage.compute_like() && !self.output,
354                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
355                    ),
356                    Bi::GlobalInvocationId
357                    | Bi::LocalInvocationId
358                    | Bi::WorkGroupId
359                    | Bi::WorkGroupSize
360                    | Bi::NumWorkGroups => (
361                        self.stage.compute_like() && !self.output,
362                        *ty_inner
363                            == Ti::Vector {
364                                size: Vs::Tri,
365                                scalar: crate::Scalar::U32,
366                            },
367                    ),
368                    Bi::NumSubgroups | Bi::SubgroupId => (
369                        self.stage.compute_like() && !self.output,
370                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
371                    ),
372                    Bi::SubgroupSize | Bi::SubgroupInvocationId => (
373                        match self.stage {
374                            St::Compute | St::Fragment | St::Task | St::Mesh => !self.output,
375                            St::Vertex => false,
376                        },
377                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
378                    ),
379                    Bi::CullPrimitive => (
380                        self.mesh_output_type == MeshOutputType::PrimitiveOutput,
381                        *ty_inner == Ti::Scalar(crate::Scalar::BOOL),
382                    ),
383                    Bi::PointIndex => (
384                        self.mesh_output_type == MeshOutputType::PrimitiveOutput,
385                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
386                    ),
387                    Bi::LineIndices => (
388                        self.mesh_output_type == MeshOutputType::PrimitiveOutput,
389                        *ty_inner
390                            == Ti::Vector {
391                                size: Vs::Bi,
392                                scalar: crate::Scalar::U32,
393                            },
394                    ),
395                    Bi::TriangleIndices => (
396                        self.mesh_output_type == MeshOutputType::PrimitiveOutput,
397                        *ty_inner
398                            == Ti::Vector {
399                                size: Vs::Tri,
400                                scalar: crate::Scalar::U32,
401                            },
402                    ),
403                    Bi::MeshTaskSize => (
404                        self.stage == St::Task && self.output,
405                        *ty_inner
406                            == Ti::Vector {
407                                size: Vs::Tri,
408                                scalar: crate::Scalar::U32,
409                            },
410                    ),
411                    // Validated elsewhere, shouldn't be here
412                    Bi::VertexCount | Bi::PrimitiveCount | Bi::Vertices | Bi::Primitives => {
413                        (false, true)
414                    }
415                };
416                match built_in {
417                    Bi::CullPrimitive
418                    | Bi::PointIndex
419                    | Bi::LineIndices
420                    | Bi::TriangleIndices
421                    | Bi::MeshTaskSize
422                    | Bi::VertexCount
423                    | Bi::PrimitiveCount
424                    | Bi::Vertices
425                    | Bi::Primitives => {
426                        if !self.capabilities.contains(Capabilities::MESH_SHADER) {
427                            return Err(VaryingError::UnsupportedCapability(
428                                Capabilities::MESH_SHADER,
429                            ));
430                        }
431                    }
432                    _ => (),
433                }
434
435                if !visible {
436                    return Err(VaryingError::InvalidBuiltInStage(built_in));
437                }
438                if !type_good {
439                    return Err(VaryingError::InvalidBuiltInType(built_in, ty_inner.clone()));
440                }
441            }
442            crate::Binding::Location {
443                location,
444                interpolation,
445                sampling,
446                blend_src,
447                per_primitive,
448            } => {
449                if per_primitive && !self.capabilities.contains(Capabilities::MESH_SHADER) {
450                    return Err(VaryingError::UnsupportedCapability(
451                        Capabilities::MESH_SHADER,
452                    ));
453                }
454                if interpolation == Some(crate::Interpolation::PerVertex) {
455                    if self.stage != crate::ShaderStage::Fragment {
456                        return Err(VaryingError::InvalidInterpolationInStage(
457                            crate::Interpolation::PerVertex,
458                            crate::ShaderStage::Fragment,
459                        ));
460                    }
461                    if !self.capabilities.contains(Capabilities::PER_VERTEX) {
462                        return Err(VaryingError::UnsupportedCapability(
463                            Capabilities::PER_VERTEX,
464                        ));
465                    }
466                }
467                // If this is per-vertex, we change the type we validate to the inner type, otherwise we leave it be.
468                // This lets all validation be done on the inner type once we've ensured the per-vertex is array<T, 3>
469                let (ty, ty_inner) = if interpolation == Some(crate::Interpolation::PerVertex) {
470                    let three = crate::ArraySize::Constant(core::num::NonZeroU32::new(3).unwrap());
471                    match ty_inner {
472                        &Ti::Array { base, size, .. } if size == three => {
473                            (base, &self.types[base].inner)
474                        }
475                        _ => return Err(VaryingError::PerVertexNotArrayOfThree),
476                    }
477                } else {
478                    (ty, ty_inner)
479                };
480
481                // Only IO-shareable types may be stored in locations.
482                if !self.type_info[ty.index()]
483                    .flags
484                    .contains(super::TypeFlags::IO_SHAREABLE)
485                {
486                    return Err(VaryingError::NotIOShareableType(ty));
487                }
488
489                // Check whether `per_primitive` is appropriate for this stage and direction.
490                if self.mesh_output_type == MeshOutputType::PrimitiveOutput {
491                    // All mesh shader `Location` outputs must be `per_primitive`.
492                    if !per_primitive {
493                        return Err(VaryingError::MissingPerPrimitive);
494                    }
495                } else if self.stage == crate::ShaderStage::Fragment && !self.output {
496                    // Fragment stage inputs may be `per_primitive`. We'll only
497                    // know if these are correct when the whole mesh pipeline is
498                    // created and we're paired with a specific mesh or vertex
499                    // shader.
500                } else if per_primitive {
501                    // All other `Location` bindings must not be `per_primitive`.
502                    return Err(VaryingError::InvalidPerPrimitive);
503                }
504
505                if let Some(blend_src) = blend_src {
506                    // `blend_src` is only valid if dual source blending was explicitly enabled,
507                    // see https://www.w3.org/TR/WGSL/#extension-dual_source_blending
508                    if !self
509                        .capabilities
510                        .contains(Capabilities::DUAL_SOURCE_BLENDING)
511                    {
512                        return Err(VaryingError::UnsupportedCapability(
513                            Capabilities::DUAL_SOURCE_BLENDING,
514                        ));
515                    }
516                    if self.stage != crate::ShaderStage::Fragment {
517                        return Err(VaryingError::InvalidAttributeInStage(
518                            "blend_src",
519                            self.stage,
520                        ));
521                    }
522                    if !self.output {
523                        return Err(VaryingError::InvalidInputAttributeInStage(
524                            "blend_src",
525                            self.stage,
526                        ));
527                    }
528                    if (blend_src != 0 && blend_src != 1) || location != 0 {
529                        return Err(VaryingError::InvalidBlendSrcIndex {
530                            location,
531                            blend_src,
532                        });
533                    }
534                    if !self.blend_src_mask.insert(blend_src as usize) {
535                        return Err(VaryingError::BindingCollisionBlendSrc { blend_src });
536                    }
537                } else if !self.location_mask.insert(location as usize)
538                    && self.flags.contains(super::ValidationFlags::BINDINGS)
539                {
540                    return Err(VaryingError::BindingCollision { location });
541                }
542
543                if let Some(interpolation) = interpolation {
544                    let invalid_sampling = match (interpolation, sampling) {
545                        (_, None)
546                        | (
547                            crate::Interpolation::Perspective | crate::Interpolation::Linear,
548                            Some(
549                                crate::Sampling::Center
550                                | crate::Sampling::Centroid
551                                | crate::Sampling::Sample,
552                            ),
553                        )
554                        | (
555                            crate::Interpolation::Flat,
556                            Some(crate::Sampling::First | crate::Sampling::Either),
557                        ) => None,
558                        (_, Some(invalid_sampling)) => Some(invalid_sampling),
559                    };
560                    if let Some(sampling) = invalid_sampling {
561                        return Err(VaryingError::InvalidInterpolationSamplingCombination {
562                            interpolation,
563                            sampling,
564                        });
565                    }
566                }
567
568                let needs_interpolation = match self.stage {
569                    crate::ShaderStage::Vertex => self.output,
570                    crate::ShaderStage::Fragment => !self.output && !per_primitive,
571                    crate::ShaderStage::Compute | crate::ShaderStage::Task => false,
572                    crate::ShaderStage::Mesh => self.output,
573                };
574
575                // It doesn't make sense to specify a sampling when `interpolation` is `Flat`, but
576                // SPIR-V and GLSL both explicitly tolerate such combinations of decorators /
577                // qualifiers, so we won't complain about that here.
578                let _ = sampling;
579
580                let required = match sampling {
581                    Some(crate::Sampling::Sample) => Capabilities::MULTISAMPLED_SHADING,
582                    _ => Capabilities::empty(),
583                };
584                if !self.capabilities.contains(required) {
585                    return Err(VaryingError::UnsupportedCapability(required));
586                }
587
588                if interpolation != Some(crate::Interpolation::PerVertex) {
589                    match ty_inner.scalar_kind() {
590                        Some(crate::ScalarKind::Float) => {
591                            if needs_interpolation && interpolation.is_none() {
592                                return Err(VaryingError::MissingInterpolation);
593                            }
594                        }
595                        Some(_) => {
596                            if needs_interpolation
597                                && interpolation != Some(crate::Interpolation::Flat)
598                            {
599                                return Err(VaryingError::InvalidInterpolation);
600                            }
601                        }
602                        None => return Err(VaryingError::InvalidType(ty)),
603                    }
604                }
605            }
606        }
607
608        Ok(())
609    }
610
611    fn validate(
612        &mut self,
613        ep: &crate::EntryPoint,
614        ty: Handle<crate::Type>,
615        binding: Option<&crate::Binding>,
616    ) -> Result<(), WithSpan<VaryingError>> {
617        let span_context = self.types.get_span_context(ty);
618        match binding {
619            Some(binding) => self
620                .validate_impl(ep, ty, binding)
621                .map_err(|e| e.with_span_context(span_context)),
622            None => {
623                let crate::TypeInner::Struct { ref members, .. } = self.types[ty].inner else {
624                    if self.flags.contains(super::ValidationFlags::BINDINGS) {
625                        return Err(VaryingError::MissingBinding.with_span());
626                    } else {
627                        return Ok(());
628                    }
629                };
630
631                for (index, member) in members.iter().enumerate() {
632                    let span_context = self.types.get_span_context(ty);
633                    match member.binding {
634                        None => {
635                            if self.flags.contains(super::ValidationFlags::BINDINGS) {
636                                return Err(VaryingError::MemberMissingBinding(index as u32)
637                                    .with_span_context(span_context));
638                            }
639                        }
640                        Some(ref binding) => self
641                            .validate_impl(ep, member.ty, binding)
642                            .map_err(|e| e.with_span_context(span_context))?,
643                    }
644                }
645
646                if !self.blend_src_mask.is_empty() {
647                    let span_context = self.types.get_span_context(ty);
648
649                    // If there's any blend_src usage, it must apply to all members of which there must be exactly 2.
650                    if members.len() != 2 || self.blend_src_mask.len() != 2 {
651                        return Err(
652                            VaryingError::IncompleteBlendSrcUsage.with_span_context(span_context)
653                        );
654                    }
655                    // Also, all members must have the same type.
656                    if members[0].ty != members[1].ty {
657                        return Err(VaryingError::BlendSrcOutputTypeMismatch {
658                            blend_src_0_type: members[0].ty,
659                            blend_src_1_type: members[1].ty,
660                        }
661                        .with_span_context(span_context));
662                    }
663                }
664                Ok(())
665            }
666        }
667    }
668}
669
670impl super::Validator {
671    pub(super) fn validate_global_var(
672        &self,
673        var: &crate::GlobalVariable,
674        gctx: crate::proc::GlobalCtx,
675        mod_info: &ModuleInfo,
676        global_expr_kind: &crate::proc::ExpressionKindTracker,
677    ) -> Result<(), GlobalVariableError> {
678        use super::TypeFlags;
679
680        log::debug!("var {var:?}");
681        let inner_ty = match gctx.types[var.ty].inner {
682            // A binding array is (mostly) supposed to behave the same as a
683            // series of individually bound resources, so we can (mostly)
684            // validate a `binding_array<T>` as if it were just a plain `T`.
685            crate::TypeInner::BindingArray { base, .. } => match var.space {
686                crate::AddressSpace::Storage { .. } => {
687                    if !self
688                        .capabilities
689                        .contains(Capabilities::STORAGE_BUFFER_BINDING_ARRAY)
690                    {
691                        return Err(GlobalVariableError::UnsupportedCapability(
692                            Capabilities::STORAGE_BUFFER_BINDING_ARRAY,
693                        ));
694                    }
695                    base
696                }
697                crate::AddressSpace::Uniform => {
698                    if !self
699                        .capabilities
700                        .contains(Capabilities::BUFFER_BINDING_ARRAY)
701                    {
702                        return Err(GlobalVariableError::UnsupportedCapability(
703                            Capabilities::BUFFER_BINDING_ARRAY,
704                        ));
705                    }
706                    base
707                }
708                crate::AddressSpace::Handle => {
709                    match gctx.types[base].inner {
710                        crate::TypeInner::Image { class, .. } => match class {
711                            crate::ImageClass::Storage { .. } => {
712                                if !self
713                                    .capabilities
714                                    .contains(Capabilities::STORAGE_TEXTURE_BINDING_ARRAY)
715                                {
716                                    return Err(GlobalVariableError::UnsupportedCapability(
717                                        Capabilities::STORAGE_TEXTURE_BINDING_ARRAY,
718                                    ));
719                                }
720                            }
721                            crate::ImageClass::Sampled { .. } | crate::ImageClass::Depth { .. } => {
722                                if !self
723                                    .capabilities
724                                    .contains(Capabilities::TEXTURE_AND_SAMPLER_BINDING_ARRAY)
725                                {
726                                    return Err(GlobalVariableError::UnsupportedCapability(
727                                        Capabilities::TEXTURE_AND_SAMPLER_BINDING_ARRAY,
728                                    ));
729                                }
730                            }
731                            crate::ImageClass::External => {
732                                // This should have been rejected in `validate_type`.
733                                unreachable!("binding arrays of external images are not supported");
734                            }
735                        },
736                        crate::TypeInner::Sampler { .. } => {
737                            if !self
738                                .capabilities
739                                .contains(Capabilities::TEXTURE_AND_SAMPLER_BINDING_ARRAY)
740                            {
741                                return Err(GlobalVariableError::UnsupportedCapability(
742                                    Capabilities::TEXTURE_AND_SAMPLER_BINDING_ARRAY,
743                                ));
744                            }
745                        }
746                        crate::TypeInner::AccelerationStructure { .. } => {
747                            return Err(GlobalVariableError::InvalidBindingArray(base));
748                        }
749                        crate::TypeInner::RayQuery { .. } => {
750                            // This should have been rejected in `validate_type`.
751                            unreachable!("binding arrays of ray queries are not supported");
752                        }
753                        _ => {
754                            // Fall through to the regular validation, which will reject `base`
755                            // as invalid in `AddressSpace::Handle`.
756                        }
757                    }
758                    base
759                }
760                _ => return Err(GlobalVariableError::InvalidUsage(var.space)),
761            },
762            _ => var.ty,
763        };
764        let type_info = &self.types[inner_ty.index()];
765
766        let (required_type_flags, is_resource) = match var.space {
767            crate::AddressSpace::Function => {
768                return Err(GlobalVariableError::InvalidUsage(var.space))
769            }
770            crate::AddressSpace::Storage { access } => {
771                if let Err((ty_handle, disalignment)) = type_info.storage_layout {
772                    if self.flags.contains(super::ValidationFlags::STRUCT_LAYOUTS) {
773                        return Err(GlobalVariableError::Alignment(
774                            var.space,
775                            ty_handle,
776                            disalignment,
777                        ));
778                    }
779                }
780                if access == crate::StorageAccess::STORE {
781                    return Err(GlobalVariableError::StorageAddressSpaceWriteOnlyNotSupported);
782                }
783                (
784                    TypeFlags::DATA | TypeFlags::HOST_SHAREABLE | TypeFlags::CREATION_RESOLVED,
785                    true,
786                )
787            }
788            crate::AddressSpace::Uniform => {
789                if let Err((ty_handle, disalignment)) = type_info.uniform_layout {
790                    if self.flags.contains(super::ValidationFlags::STRUCT_LAYOUTS) {
791                        return Err(GlobalVariableError::Alignment(
792                            var.space,
793                            ty_handle,
794                            disalignment,
795                        ));
796                    }
797                }
798                (
799                    TypeFlags::DATA
800                        | TypeFlags::COPY
801                        | TypeFlags::SIZED
802                        | TypeFlags::HOST_SHAREABLE
803                        | TypeFlags::CREATION_RESOLVED,
804                    true,
805                )
806            }
807            crate::AddressSpace::Handle => {
808                match gctx.types[inner_ty].inner {
809                    crate::TypeInner::Image { class, .. } => match class {
810                        crate::ImageClass::Storage {
811                            format:
812                                crate::StorageFormat::R16Unorm
813                                | crate::StorageFormat::R16Snorm
814                                | crate::StorageFormat::Rg16Unorm
815                                | crate::StorageFormat::Rg16Snorm
816                                | crate::StorageFormat::Rgba16Unorm
817                                | crate::StorageFormat::Rgba16Snorm,
818                            ..
819                        } => {
820                            if !self
821                                .capabilities
822                                .contains(Capabilities::STORAGE_TEXTURE_16BIT_NORM_FORMATS)
823                            {
824                                return Err(GlobalVariableError::UnsupportedCapability(
825                                    Capabilities::STORAGE_TEXTURE_16BIT_NORM_FORMATS,
826                                ));
827                            }
828                        }
829                        _ => {}
830                    },
831                    crate::TypeInner::Sampler { .. }
832                    | crate::TypeInner::AccelerationStructure { .. }
833                    | crate::TypeInner::RayQuery { .. } => {}
834                    _ => {
835                        return Err(GlobalVariableError::InvalidType(var.space));
836                    }
837                }
838
839                (TypeFlags::empty(), true)
840            }
841            crate::AddressSpace::Private => (
842                TypeFlags::CONSTRUCTIBLE | TypeFlags::CREATION_RESOLVED,
843                false,
844            ),
845            crate::AddressSpace::WorkGroup => (TypeFlags::DATA | TypeFlags::SIZED, false),
846            crate::AddressSpace::TaskPayload => {
847                if !self.capabilities.contains(Capabilities::MESH_SHADER) {
848                    return Err(GlobalVariableError::UnsupportedCapability(
849                        Capabilities::MESH_SHADER,
850                    ));
851                }
852                (TypeFlags::DATA | TypeFlags::SIZED, false)
853            }
854            crate::AddressSpace::Immediate => {
855                if !self.capabilities.contains(Capabilities::IMMEDIATES) {
856                    return Err(GlobalVariableError::UnsupportedCapability(
857                        Capabilities::IMMEDIATES,
858                    ));
859                }
860                if let Err(ref err) = type_info.immediates_compatibility {
861                    return Err(GlobalVariableError::InvalidImmediateType(err.clone()));
862                }
863                (
864                    TypeFlags::DATA
865                        | TypeFlags::COPY
866                        | TypeFlags::HOST_SHAREABLE
867                        | TypeFlags::SIZED,
868                    false,
869                )
870            }
871        };
872
873        if !type_info.flags.contains(required_type_flags) {
874            return Err(GlobalVariableError::MissingTypeFlags {
875                seen: type_info.flags,
876                required: required_type_flags,
877            });
878        }
879
880        if is_resource != var.binding.is_some() {
881            if self.flags.contains(super::ValidationFlags::BINDINGS) {
882                return Err(GlobalVariableError::InvalidBinding);
883            }
884        }
885
886        if var.space == crate::AddressSpace::TaskPayload {
887            let ty = &gctx.types[var.ty].inner;
888            // HLSL doesn't allow zero sized payloads.
889            if ty.try_size(gctx) == Some(0) {
890                return Err(GlobalVariableError::ZeroSizedTaskPayload);
891            }
892        }
893
894        if let Some(init) = var.init {
895            match var.space {
896                crate::AddressSpace::Private | crate::AddressSpace::Function => {}
897                _ => {
898                    return Err(GlobalVariableError::InitializerNotAllowed(var.space));
899                }
900            }
901
902            if !global_expr_kind.is_const_or_override(init) {
903                return Err(GlobalVariableError::InitializerExprType);
904            }
905
906            if !gctx.compare_types(
907                &crate::proc::TypeResolution::Handle(var.ty),
908                &mod_info[init],
909            ) {
910                return Err(GlobalVariableError::InitializerType);
911            }
912        }
913
914        Ok(())
915    }
916
917    /// Validate the mesh shader output type `ty`, used as `mesh_output_type`.
918    fn validate_mesh_output_type(
919        &mut self,
920        ep: &crate::EntryPoint,
921        module: &crate::Module,
922        ty: Handle<crate::Type>,
923        mesh_output_type: MeshOutputType,
924    ) -> Result<(), WithSpan<EntryPointError>> {
925        if !matches!(module.types[ty].inner, crate::TypeInner::Struct { .. }) {
926            return Err(EntryPointError::InvalidMeshOutputType.with_span_handle(ty, &module.types));
927        }
928        let mut result_built_ins = crate::FastHashSet::default();
929        let mut ctx = VaryingContext {
930            stage: ep.stage,
931            output: true,
932            types: &module.types,
933            type_info: &self.types,
934            location_mask: &mut self.location_mask,
935            blend_src_mask: &mut self.blend_src_mask,
936            built_ins: &mut result_built_ins,
937            capabilities: self.capabilities,
938            flags: self.flags,
939            mesh_output_type,
940            has_task_payload: ep.task_payload.is_some(),
941        };
942        ctx.validate(ep, ty, None)
943            .map_err_inner(|e| EntryPointError::Result(e).with_span())?;
944        if mesh_output_type == MeshOutputType::PrimitiveOutput {
945            let mut num_indices_builtins = 0;
946            if result_built_ins.contains(&crate::BuiltIn::PointIndex) {
947                num_indices_builtins += 1;
948            }
949            if result_built_ins.contains(&crate::BuiltIn::LineIndices) {
950                num_indices_builtins += 1;
951            }
952            if result_built_ins.contains(&crate::BuiltIn::TriangleIndices) {
953                num_indices_builtins += 1;
954            }
955            if num_indices_builtins != 1 {
956                return Err(EntryPointError::InvalidMeshPrimitiveOutputType
957                    .with_span_handle(ty, &module.types));
958            }
959        } else if mesh_output_type == MeshOutputType::VertexOutput
960            && !result_built_ins.contains(&crate::BuiltIn::Position { invariant: false })
961        {
962            return Err(
963                EntryPointError::MissingVertexOutputPosition.with_span_handle(ty, &module.types)
964            );
965        }
966
967        Ok(())
968    }
969
970    pub(super) fn validate_entry_point(
971        &mut self,
972        ep: &crate::EntryPoint,
973        module: &crate::Module,
974        mod_info: &ModuleInfo,
975    ) -> Result<FunctionInfo, WithSpan<EntryPointError>> {
976        if matches!(
977            ep.stage,
978            crate::ShaderStage::Task | crate::ShaderStage::Mesh
979        ) && !self.capabilities.contains(Capabilities::MESH_SHADER)
980        {
981            return Err(
982                EntryPointError::UnsupportedCapability(Capabilities::MESH_SHADER).with_span(),
983            );
984        }
985        if ep.early_depth_test.is_some() {
986            let required = Capabilities::EARLY_DEPTH_TEST;
987            if !self.capabilities.contains(required) {
988                return Err(
989                    EntryPointError::Result(VaryingError::UnsupportedCapability(required))
990                        .with_span(),
991                );
992            }
993
994            if ep.stage != crate::ShaderStage::Fragment {
995                return Err(EntryPointError::UnexpectedEarlyDepthTest.with_span());
996            }
997        }
998
999        if ep.stage.compute_like() {
1000            if ep
1001                .workgroup_size
1002                .iter()
1003                .any(|&s| s == 0 || s > MAX_WORKGROUP_SIZE)
1004            {
1005                return Err(EntryPointError::OutOfRangeWorkgroupSize.with_span());
1006            }
1007        } else if ep.workgroup_size != [0; 3] {
1008            return Err(EntryPointError::UnexpectedWorkgroupSize.with_span());
1009        }
1010
1011        match (ep.stage, &ep.mesh_info) {
1012            (crate::ShaderStage::Mesh, &None) => {
1013                return Err(EntryPointError::ExpectedMeshShaderAttributes.with_span());
1014            }
1015            (crate::ShaderStage::Mesh, &Some(..)) => {}
1016            (_, &Some(_)) => {
1017                return Err(EntryPointError::UnexpectedMeshShaderAttributes.with_span());
1018            }
1019            (_, _) => {}
1020        }
1021
1022        let mut info = self
1023            .validate_function(&ep.function, module, mod_info, true)
1024            .map_err(WithSpan::into_other)?;
1025
1026        // Validate the task shader payload.
1027        match ep.stage {
1028            // Task shaders must produce a payload.
1029            crate::ShaderStage::Task => {
1030                let Some(handle) = ep.task_payload else {
1031                    return Err(EntryPointError::ExpectedTaskPayload.with_span());
1032                };
1033                if module.global_variables[handle].space != crate::AddressSpace::TaskPayload {
1034                    return Err(EntryPointError::TaskPayloadWrongAddressSpace
1035                        .with_span_handle(handle, &module.global_variables));
1036                }
1037                info.insert_global_use(GlobalUse::READ | GlobalUse::WRITE, handle);
1038            }
1039
1040            // Mesh shaders may accept a payload.
1041            crate::ShaderStage::Mesh => {
1042                if let Some(handle) = ep.task_payload {
1043                    if module.global_variables[handle].space != crate::AddressSpace::TaskPayload {
1044                        return Err(EntryPointError::TaskPayloadWrongAddressSpace
1045                            .with_span_handle(handle, &module.global_variables));
1046                    }
1047                    info.insert_global_use(GlobalUse::READ, handle);
1048                }
1049                if let Some(ref mesh_info) = ep.mesh_info {
1050                    info.insert_global_use(GlobalUse::READ, mesh_info.output_variable);
1051                }
1052            }
1053
1054            // Other stages must not have a payload.
1055            _ => {
1056                if let Some(handle) = ep.task_payload {
1057                    return Err(EntryPointError::UnexpectedTaskPayload
1058                        .with_span_handle(handle, &module.global_variables));
1059                }
1060            }
1061        }
1062
1063        {
1064            use super::ShaderStages;
1065
1066            let stage_bit = match ep.stage {
1067                crate::ShaderStage::Vertex => ShaderStages::VERTEX,
1068                crate::ShaderStage::Fragment => ShaderStages::FRAGMENT,
1069                crate::ShaderStage::Compute => ShaderStages::COMPUTE,
1070                crate::ShaderStage::Mesh => ShaderStages::MESH,
1071                crate::ShaderStage::Task => ShaderStages::TASK,
1072            };
1073
1074            if !info.available_stages.contains(stage_bit) {
1075                return Err(EntryPointError::ForbiddenStageOperations.with_span());
1076            }
1077        }
1078
1079        self.location_mask.clear();
1080        let mut argument_built_ins = crate::FastHashSet::default();
1081        // TODO: add span info to function arguments
1082        for (index, fa) in ep.function.arguments.iter().enumerate() {
1083            let mut ctx = VaryingContext {
1084                stage: ep.stage,
1085                output: false,
1086                types: &module.types,
1087                type_info: &self.types,
1088                location_mask: &mut self.location_mask,
1089                blend_src_mask: &mut self.blend_src_mask,
1090                built_ins: &mut argument_built_ins,
1091                capabilities: self.capabilities,
1092                flags: self.flags,
1093                mesh_output_type: MeshOutputType::None,
1094                has_task_payload: ep.task_payload.is_some(),
1095            };
1096            ctx.validate(ep, fa.ty, fa.binding.as_ref())
1097                .map_err_inner(|e| EntryPointError::Argument(index as u32, e).with_span())?;
1098        }
1099
1100        self.location_mask.clear();
1101        if let Some(ref fr) = ep.function.result {
1102            let mut result_built_ins = crate::FastHashSet::default();
1103            let mut ctx = VaryingContext {
1104                stage: ep.stage,
1105                output: true,
1106                types: &module.types,
1107                type_info: &self.types,
1108                location_mask: &mut self.location_mask,
1109                blend_src_mask: &mut self.blend_src_mask,
1110                built_ins: &mut result_built_ins,
1111                capabilities: self.capabilities,
1112                flags: self.flags,
1113                mesh_output_type: MeshOutputType::None,
1114                has_task_payload: ep.task_payload.is_some(),
1115            };
1116            ctx.validate(ep, fr.ty, fr.binding.as_ref())
1117                .map_err_inner(|e| EntryPointError::Result(e).with_span())?;
1118            if ep.stage == crate::ShaderStage::Vertex
1119                && !result_built_ins.contains(&crate::BuiltIn::Position { invariant: false })
1120            {
1121                return Err(EntryPointError::MissingVertexOutputPosition.with_span());
1122            }
1123            if ep.stage == crate::ShaderStage::Mesh {
1124                return Err(EntryPointError::UnexpectedMeshShaderEntryResult.with_span());
1125            }
1126            // Task shaders must have a single `MeshTaskSize` output, and nothing else.
1127            if ep.stage == crate::ShaderStage::Task {
1128                let ok = result_built_ins.contains(&crate::BuiltIn::MeshTaskSize)
1129                    && result_built_ins.len() == 1
1130                    && self.location_mask.is_empty();
1131                if !ok {
1132                    return Err(EntryPointError::WrongTaskShaderEntryResult.with_span());
1133                }
1134            }
1135            if !self.blend_src_mask.is_empty() {
1136                info.dual_source_blending = true;
1137            }
1138        } else if ep.stage == crate::ShaderStage::Vertex {
1139            return Err(EntryPointError::MissingVertexOutputPosition.with_span());
1140        } else if ep.stage == crate::ShaderStage::Task {
1141            return Err(EntryPointError::WrongTaskShaderEntryResult.with_span());
1142        }
1143
1144        {
1145            let mut used_immediates = module
1146                .global_variables
1147                .iter()
1148                .filter(|&(_, var)| var.space == crate::AddressSpace::Immediate)
1149                .map(|(handle, _)| handle)
1150                .filter(|&handle| !info[handle].is_empty());
1151            // Check if there is more than one immediate data, and error if so.
1152            // Use a loop for when returning multiple errors is supported.
1153            if let Some(handle) = used_immediates.nth(1) {
1154                return Err(EntryPointError::MoreThanOneImmediateUsed
1155                    .with_span_handle(handle, &module.global_variables));
1156            }
1157        }
1158
1159        self.ep_resource_bindings.clear();
1160        for (var_handle, var) in module.global_variables.iter() {
1161            let usage = info[var_handle];
1162            if usage.is_empty() {
1163                continue;
1164            }
1165
1166            if var.space == crate::AddressSpace::TaskPayload {
1167                if ep.task_payload != Some(var_handle) {
1168                    return Err(EntryPointError::WrongTaskPayloadUsed
1169                        .with_span_handle(var_handle, &module.global_variables));
1170                }
1171                let size = module.types[var.ty].inner.size(module.to_ctx());
1172                if size < 4 {
1173                    return Err(EntryPointError::TaskPayloadTooSmall(size)
1174                        .with_span_handle(var_handle, &module.global_variables));
1175                }
1176            }
1177
1178            let allowed_usage = match var.space {
1179                crate::AddressSpace::Function => unreachable!(),
1180                crate::AddressSpace::Uniform => GlobalUse::READ | GlobalUse::QUERY,
1181                crate::AddressSpace::Storage { access } => storage_usage(access),
1182                crate::AddressSpace::Handle => match module.types[var.ty].inner {
1183                    crate::TypeInner::BindingArray { base, .. } => match module.types[base].inner {
1184                        crate::TypeInner::Image {
1185                            class: crate::ImageClass::Storage { access, .. },
1186                            ..
1187                        } => storage_usage(access),
1188                        _ => GlobalUse::READ | GlobalUse::QUERY,
1189                    },
1190                    crate::TypeInner::Image {
1191                        class: crate::ImageClass::Storage { access, .. },
1192                        ..
1193                    } => storage_usage(access),
1194                    _ => GlobalUse::READ | GlobalUse::QUERY,
1195                },
1196                crate::AddressSpace::Private | crate::AddressSpace::WorkGroup => {
1197                    GlobalUse::READ | GlobalUse::WRITE | GlobalUse::QUERY
1198                }
1199                crate::AddressSpace::TaskPayload => {
1200                    GlobalUse::READ
1201                        | GlobalUse::QUERY
1202                        | if ep.stage == crate::ShaderStage::Task {
1203                            GlobalUse::WRITE
1204                        } else {
1205                            GlobalUse::empty()
1206                        }
1207                }
1208                crate::AddressSpace::Immediate => GlobalUse::READ,
1209            };
1210            if !allowed_usage.contains(usage) {
1211                log::warn!("\tUsage error for: {var:?}");
1212                log::warn!("\tAllowed usage: {allowed_usage:?}, requested: {usage:?}");
1213                return Err(EntryPointError::InvalidGlobalUsage(var_handle, usage)
1214                    .with_span_handle(var_handle, &module.global_variables));
1215            }
1216
1217            if let Some(ref bind) = var.binding {
1218                if !self.ep_resource_bindings.insert(*bind) {
1219                    if self.flags.contains(super::ValidationFlags::BINDINGS) {
1220                        return Err(EntryPointError::BindingCollision(var_handle)
1221                            .with_span_handle(var_handle, &module.global_variables));
1222                    }
1223                }
1224            }
1225        }
1226
1227        // If this is a `Mesh` entry point, check its vertex and primitive output types.
1228        // We verified previously that only mesh shaders can have `mesh_info`.
1229        if let &Some(ref mesh_info) = &ep.mesh_info {
1230            if module.global_variables[mesh_info.output_variable].space
1231                != crate::AddressSpace::WorkGroup
1232            {
1233                return Err(EntryPointError::WrongMeshOutputAddressSpace.with_span());
1234            }
1235
1236            let mut implied = module.analyze_mesh_shader_info(mesh_info.output_variable);
1237            if let Some(e) = implied.2 {
1238                return Err(e);
1239            }
1240
1241            if let Some(e) = mesh_info.max_vertices_override {
1242                if let crate::Expression::Override(o) = module.global_expressions[e] {
1243                    if implied.1[0] != Some(o) {
1244                        return Err(EntryPointError::BadMeshOutputVariableType.with_span());
1245                    }
1246                }
1247            }
1248            if let Some(e) = mesh_info.max_primitives_override {
1249                if let crate::Expression::Override(o) = module.global_expressions[e] {
1250                    if implied.1[1] != Some(o) {
1251                        return Err(EntryPointError::BadMeshOutputVariableType.with_span());
1252                    }
1253                }
1254            }
1255
1256            implied.0.max_vertices_override = mesh_info.max_vertices_override;
1257            implied.0.max_primitives_override = mesh_info.max_primitives_override;
1258            if implied.0 != *mesh_info {
1259                return Err(EntryPointError::BadMeshOutputVariableType.with_span());
1260            }
1261            if mesh_info.topology == crate::MeshOutputTopology::Points
1262                && !self
1263                    .capabilities
1264                    .contains(Capabilities::MESH_SHADER_POINT_TOPOLOGY)
1265            {
1266                return Err(EntryPointError::UnsupportedCapability(
1267                    Capabilities::MESH_SHADER_POINT_TOPOLOGY,
1268                )
1269                .with_span());
1270            }
1271
1272            self.validate_mesh_output_type(
1273                ep,
1274                module,
1275                mesh_info.vertex_output_type,
1276                MeshOutputType::VertexOutput,
1277            )?;
1278            self.validate_mesh_output_type(
1279                ep,
1280                module,
1281                mesh_info.primitive_output_type,
1282                MeshOutputType::PrimitiveOutput,
1283            )?;
1284        }
1285
1286        Ok(info)
1287    }
1288}
naga/valid/interface.rs

naga/valid/
interface.rs
