naga/valid/
interface.rs

Help
1use alloc::vec::Vec;
2
3use bit_set::BitSet;
4
5use super::{
6    analyzer::{FunctionInfo, GlobalUse},
7    Capabilities, Disalignment, FunctionError, ModuleInfo, PushConstantError,
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 flags {seen:?} do not meet the required {required:?}")]
22    MissingTypeFlags {
23        required: super::TypeFlags,
24        seen: super::TypeFlags,
25    },
26    #[error("Capability {0:?} is not supported")]
27    UnsupportedCapability(Capabilities),
28    #[error("Binding decoration is missing or not applicable")]
29    InvalidBinding,
30    #[error("Alignment requirements for address space {0:?} are not met by {1:?}")]
31    Alignment(
32        crate::AddressSpace,
33        Handle<crate::Type>,
34        #[source] Disalignment,
35    ),
36    #[error("Initializer must be an override-expression")]
37    InitializerExprType,
38    #[error("Initializer doesn't match the variable type")]
39    InitializerType,
40    #[error("Initializer can't be used with address space {0:?}")]
41    InitializerNotAllowed(crate::AddressSpace),
42    #[error("Storage address space doesn't support write-only access")]
43    StorageAddressSpaceWriteOnlyNotSupported,
44    #[error("Type is not valid for use as a push constant")]
45    InvalidPushConstantType(#[source] PushConstantError),
46}
47
48#[derive(Clone, Debug, thiserror::Error)]
49#[cfg_attr(test, derive(PartialEq))]
50pub enum VaryingError {
51    #[error("The type {0:?} does not match the varying")]
52    InvalidType(Handle<crate::Type>),
53    #[error("The type {0:?} cannot be used for user-defined entry point inputs or outputs")]
54    NotIOShareableType(Handle<crate::Type>),
55    #[error("Interpolation is not valid")]
56    InvalidInterpolation,
57    #[error("Cannot combine {interpolation:?} interpolation with the {sampling:?} sample type")]
58    InvalidInterpolationSamplingCombination {
59        interpolation: crate::Interpolation,
60        sampling: crate::Sampling,
61    },
62    #[error("Interpolation must be specified on vertex shader outputs and fragment shader inputs")]
63    MissingInterpolation,
64    #[error("Built-in {0:?} is not available at this stage")]
65    InvalidBuiltInStage(crate::BuiltIn),
66    #[error("Built-in type for {0:?} is invalid")]
67    InvalidBuiltInType(crate::BuiltIn),
68    #[error("Entry point arguments and return values must all have bindings")]
69    MissingBinding,
70    #[error("Struct member {0} is missing a binding")]
71    MemberMissingBinding(u32),
72    #[error("Multiple bindings at location {location} are present")]
73    BindingCollision { location: u32 },
74    #[error("Multiple bindings use the same `blend_src` {blend_src}")]
75    BindingCollisionBlendSrc { blend_src: u32 },
76    #[error("Built-in {0:?} is present more than once")]
77    DuplicateBuiltIn(crate::BuiltIn),
78    #[error("Capability {0:?} is not supported")]
79    UnsupportedCapability(Capabilities),
80    #[error("The attribute {0:?} is only valid as an output for stage {1:?}")]
81    InvalidInputAttributeInStage(&'static str, crate::ShaderStage),
82    #[error("The attribute {0:?} is not valid for stage {1:?}")]
83    InvalidAttributeInStage(&'static str, crate::ShaderStage),
84    #[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}.")]
85    InvalidBlendSrcIndex { location: u32, blend_src: u32 },
86    #[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)`.")]
87    IncompleteBlendSrcUsage,
88    #[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:?}.")]
89    BlendSrcOutputTypeMismatch {
90        blend_src_0_type: Handle<crate::Type>,
91        blend_src_1_type: Handle<crate::Type>,
92    },
93    #[error("Workgroup size is multi dimensional, `@builtin(subgroup_id)` and `@builtin(subgroup_invocation_id)` are not supported.")]
94    InvalidMultiDimensionalSubgroupBuiltIn,
95}
96
97#[derive(Clone, Debug, thiserror::Error)]
98#[cfg_attr(test, derive(PartialEq))]
99pub enum EntryPointError {
100    #[error("Multiple conflicting entry points")]
101    Conflict,
102    #[error("Vertex shaders must return a `@builtin(position)` output value")]
103    MissingVertexOutputPosition,
104    #[error("Early depth test is not applicable")]
105    UnexpectedEarlyDepthTest,
106    #[error("Workgroup size is not applicable")]
107    UnexpectedWorkgroupSize,
108    #[error("Workgroup size is out of range")]
109    OutOfRangeWorkgroupSize,
110    #[error("Uses operations forbidden at this stage")]
111    ForbiddenStageOperations,
112    #[error("Global variable {0:?} is used incorrectly as {1:?}")]
113    InvalidGlobalUsage(Handle<crate::GlobalVariable>, GlobalUse),
114    #[error("More than 1 push constant variable is used")]
115    MoreThanOnePushConstantUsed,
116    #[error("Bindings for {0:?} conflict with other resource")]
117    BindingCollision(Handle<crate::GlobalVariable>),
118    #[error("Argument {0} varying error")]
119    Argument(u32, #[source] VaryingError),
120    #[error(transparent)]
121    Result(#[from] VaryingError),
122    #[error("Location {location} interpolation of an integer has to be flat")]
123    InvalidIntegerInterpolation { location: u32 },
124    #[error(transparent)]
125    Function(#[from] FunctionError),
126}
127
128fn storage_usage(access: crate::StorageAccess) -> GlobalUse {
129    let mut storage_usage = GlobalUse::QUERY;
130    if access.contains(crate::StorageAccess::LOAD) {
131        storage_usage |= GlobalUse::READ;
132    }
133    if access.contains(crate::StorageAccess::STORE) {
134        storage_usage |= GlobalUse::WRITE;
135    }
136    if access.contains(crate::StorageAccess::ATOMIC) {
137        storage_usage |= GlobalUse::ATOMIC;
138    }
139    storage_usage
140}
141
142struct VaryingContext<'a> {
143    stage: crate::ShaderStage,
144    output: bool,
145    types: &'a UniqueArena<crate::Type>,
146    type_info: &'a Vec<super::r#type::TypeInfo>,
147    location_mask: &'a mut BitSet,
148    blend_src_mask: &'a mut BitSet,
149    built_ins: &'a mut crate::FastHashSet<crate::BuiltIn>,
150    capabilities: Capabilities,
151    flags: super::ValidationFlags,
152}
153
154impl VaryingContext<'_> {
155    fn validate_impl(
156        &mut self,
157        ep: &crate::EntryPoint,
158        ty: Handle<crate::Type>,
159        binding: &crate::Binding,
160    ) -> Result<(), VaryingError> {
161        use crate::{BuiltIn as Bi, ShaderStage as St, TypeInner as Ti, VectorSize as Vs};
162
163        let ty_inner = &self.types[ty].inner;
164        match *binding {
165            crate::Binding::BuiltIn(built_in) => {
166                // Ignore the `invariant` field for the sake of duplicate checks,
167                // but use the original in error messages.
168                let canonical = if let crate::BuiltIn::Position { .. } = built_in {
169                    crate::BuiltIn::Position { invariant: false }
170                } else {
171                    built_in
172                };
173
174                if self.built_ins.contains(&canonical) {
175                    return Err(VaryingError::DuplicateBuiltIn(built_in));
176                }
177                self.built_ins.insert(canonical);
178
179                let required = match built_in {
180                    Bi::ClipDistance => Capabilities::CLIP_DISTANCE,
181                    Bi::CullDistance => Capabilities::CULL_DISTANCE,
182                    Bi::PrimitiveIndex => Capabilities::PRIMITIVE_INDEX,
183                    Bi::ViewIndex => Capabilities::MULTIVIEW,
184                    Bi::SampleIndex => Capabilities::MULTISAMPLED_SHADING,
185                    Bi::NumSubgroups
186                    | Bi::SubgroupId
187                    | Bi::SubgroupSize
188                    | Bi::SubgroupInvocationId => Capabilities::SUBGROUP,
189                    _ => Capabilities::empty(),
190                };
191                if !self.capabilities.contains(required) {
192                    return Err(VaryingError::UnsupportedCapability(required));
193                }
194
195                if matches!(
196                    built_in,
197                    crate::BuiltIn::SubgroupId | crate::BuiltIn::SubgroupInvocationId
198                ) && ep.workgroup_size[1..].iter().any(|&s| s > 1)
199                {
200                    return Err(VaryingError::InvalidMultiDimensionalSubgroupBuiltIn);
201                }
202
203                let (visible, type_good) = match built_in {
204                    Bi::BaseInstance
205                    | Bi::BaseVertex
206                    | Bi::InstanceIndex
207                    | Bi::VertexIndex
208                    | Bi::DrawID => (
209                        self.stage == St::Vertex && !self.output,
210                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
211                    ),
212                    Bi::ClipDistance | Bi::CullDistance => (
213                        self.stage == St::Vertex && self.output,
214                        match *ty_inner {
215                            Ti::Array { base, size, .. } => {
216                                self.types[base].inner == Ti::Scalar(crate::Scalar::F32)
217                                    && match size {
218                                        crate::ArraySize::Constant(non_zero) => non_zero.get() <= 8,
219                                        _ => false,
220                                    }
221                            }
222                            _ => false,
223                        },
224                    ),
225                    Bi::PointSize => (
226                        self.stage == St::Vertex && self.output,
227                        *ty_inner == Ti::Scalar(crate::Scalar::F32),
228                    ),
229                    Bi::PointCoord => (
230                        self.stage == St::Fragment && !self.output,
231                        *ty_inner
232                            == Ti::Vector {
233                                size: Vs::Bi,
234                                scalar: crate::Scalar::F32,
235                            },
236                    ),
237                    Bi::Position { .. } => (
238                        match self.stage {
239                            St::Vertex => self.output,
240                            St::Fragment => !self.output,
241                            St::Compute => false,
242                            St::Task | St::Mesh => unreachable!(),
243                        },
244                        *ty_inner
245                            == Ti::Vector {
246                                size: Vs::Quad,
247                                scalar: crate::Scalar::F32,
248                            },
249                    ),
250                    Bi::ViewIndex => (
251                        match self.stage {
252                            St::Vertex | St::Fragment => !self.output,
253                            St::Compute => false,
254                            St::Task | St::Mesh => unreachable!(),
255                        },
256                        *ty_inner == Ti::Scalar(crate::Scalar::I32),
257                    ),
258                    Bi::FragDepth => (
259                        self.stage == St::Fragment && self.output,
260                        *ty_inner == Ti::Scalar(crate::Scalar::F32),
261                    ),
262                    Bi::FrontFacing => (
263                        self.stage == St::Fragment && !self.output,
264                        *ty_inner == Ti::Scalar(crate::Scalar::BOOL),
265                    ),
266                    Bi::PrimitiveIndex => (
267                        self.stage == St::Fragment && !self.output,
268                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
269                    ),
270                    Bi::SampleIndex => (
271                        self.stage == St::Fragment && !self.output,
272                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
273                    ),
274                    Bi::SampleMask => (
275                        self.stage == St::Fragment,
276                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
277                    ),
278                    Bi::LocalInvocationIndex => (
279                        self.stage == St::Compute && !self.output,
280                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
281                    ),
282                    Bi::GlobalInvocationId
283                    | Bi::LocalInvocationId
284                    | Bi::WorkGroupId
285                    | Bi::WorkGroupSize
286                    | Bi::NumWorkGroups => (
287                        self.stage == St::Compute && !self.output,
288                        *ty_inner
289                            == Ti::Vector {
290                                size: Vs::Tri,
291                                scalar: crate::Scalar::U32,
292                            },
293                    ),
294                    Bi::NumSubgroups | Bi::SubgroupId => (
295                        self.stage == St::Compute && !self.output,
296                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
297                    ),
298                    Bi::SubgroupSize | Bi::SubgroupInvocationId => (
299                        match self.stage {
300                            St::Compute | St::Fragment => !self.output,
301                            St::Vertex => false,
302                            St::Task | St::Mesh => unreachable!(),
303                        },
304                        *ty_inner == Ti::Scalar(crate::Scalar::U32),
305                    ),
306                };
307
308                if !visible {
309                    return Err(VaryingError::InvalidBuiltInStage(built_in));
310                }
311                if !type_good {
312                    log::warn!("Wrong builtin type: {ty_inner:?}");
313                    return Err(VaryingError::InvalidBuiltInType(built_in));
314                }
315            }
316            crate::Binding::Location {
317                location,
318                interpolation,
319                sampling,
320                blend_src,
321            } => {
322                // Only IO-shareable types may be stored in locations.
323                if !self.type_info[ty.index()]
324                    .flags
325                    .contains(super::TypeFlags::IO_SHAREABLE)
326                {
327                    return Err(VaryingError::NotIOShareableType(ty));
328                }
329
330                if let Some(blend_src) = blend_src {
331                    // `blend_src` is only valid if dual source blending was explicitly enabled,
332                    // see https://www.w3.org/TR/WGSL/#extension-dual_source_blending
333                    if !self
334                        .capabilities
335                        .contains(Capabilities::DUAL_SOURCE_BLENDING)
336                    {
337                        return Err(VaryingError::UnsupportedCapability(
338                            Capabilities::DUAL_SOURCE_BLENDING,
339                        ));
340                    }
341                    if self.stage != crate::ShaderStage::Fragment {
342                        return Err(VaryingError::InvalidAttributeInStage(
343                            "blend_src",
344                            self.stage,
345                        ));
346                    }
347                    if !self.output {
348                        return Err(VaryingError::InvalidInputAttributeInStage(
349                            "blend_src",
350                            self.stage,
351                        ));
352                    }
353                    if (blend_src != 0 && blend_src != 1) || location != 0 {
354                        return Err(VaryingError::InvalidBlendSrcIndex {
355                            location,
356                            blend_src,
357                        });
358                    }
359                    if !self.blend_src_mask.insert(blend_src as usize) {
360                        return Err(VaryingError::BindingCollisionBlendSrc { blend_src });
361                    }
362                } else if !self.location_mask.insert(location as usize)
363                    && self.flags.contains(super::ValidationFlags::BINDINGS)
364                {
365                    return Err(VaryingError::BindingCollision { location });
366                }
367
368                if let Some(interpolation) = interpolation {
369                    let invalid_sampling = match (interpolation, sampling) {
370                        (_, None)
371                        | (
372                            crate::Interpolation::Perspective | crate::Interpolation::Linear,
373                            Some(
374                                crate::Sampling::Center
375                                | crate::Sampling::Centroid
376                                | crate::Sampling::Sample,
377                            ),
378                        )
379                        | (
380                            crate::Interpolation::Flat,
381                            Some(crate::Sampling::First | crate::Sampling::Either),
382                        ) => None,
383                        (_, Some(invalid_sampling)) => Some(invalid_sampling),
384                    };
385                    if let Some(sampling) = invalid_sampling {
386                        return Err(VaryingError::InvalidInterpolationSamplingCombination {
387                            interpolation,
388                            sampling,
389                        });
390                    }
391                }
392
393                let needs_interpolation = match self.stage {
394                    crate::ShaderStage::Vertex => self.output,
395                    crate::ShaderStage::Fragment => !self.output,
396                    crate::ShaderStage::Compute => false,
397                    crate::ShaderStage::Task | crate::ShaderStage::Mesh => unreachable!(),
398                };
399
400                // It doesn't make sense to specify a sampling when `interpolation` is `Flat`, but
401                // SPIR-V and GLSL both explicitly tolerate such combinations of decorators /
402                // qualifiers, so we won't complain about that here.
403                let _ = sampling;
404
405                let required = match sampling {
406                    Some(crate::Sampling::Sample) => Capabilities::MULTISAMPLED_SHADING,
407                    _ => Capabilities::empty(),
408                };
409                if !self.capabilities.contains(required) {
410                    return Err(VaryingError::UnsupportedCapability(required));
411                }
412
413                match ty_inner.scalar_kind() {
414                    Some(crate::ScalarKind::Float) => {
415                        if needs_interpolation && interpolation.is_none() {
416                            return Err(VaryingError::MissingInterpolation);
417                        }
418                    }
419                    Some(_) => {
420                        if needs_interpolation && interpolation != Some(crate::Interpolation::Flat)
421                        {
422                            return Err(VaryingError::InvalidInterpolation);
423                        }
424                    }
425                    None => return Err(VaryingError::InvalidType(ty)),
426                }
427            }
428        }
429
430        Ok(())
431    }
432
433    fn validate(
434        &mut self,
435        ep: &crate::EntryPoint,
436        ty: Handle<crate::Type>,
437        binding: Option<&crate::Binding>,
438    ) -> Result<(), WithSpan<VaryingError>> {
439        let span_context = self.types.get_span_context(ty);
440        match binding {
441            Some(binding) => self
442                .validate_impl(ep, ty, binding)
443                .map_err(|e| e.with_span_context(span_context)),
444            None => {
445                match self.types[ty].inner {
446                    crate::TypeInner::Struct { ref members, .. } => {
447                        for (index, member) in members.iter().enumerate() {
448                            let span_context = self.types.get_span_context(ty);
449                            match member.binding {
450                                None => {
451                                    if self.flags.contains(super::ValidationFlags::BINDINGS) {
452                                        return Err(VaryingError::MemberMissingBinding(
453                                            index as u32,
454                                        )
455                                        .with_span_context(span_context));
456                                    }
457                                }
458                                Some(ref binding) => self
459                                    .validate_impl(ep, member.ty, binding)
460                                    .map_err(|e| e.with_span_context(span_context))?,
461                            }
462                        }
463
464                        if !self.blend_src_mask.is_empty() {
465                            let span_context = self.types.get_span_context(ty);
466
467                            // If there's any blend_src usage, it must apply to all members of which there must be exactly 2.
468                            if members.len() != 2 || self.blend_src_mask.len() != 2 {
469                                return Err(VaryingError::IncompleteBlendSrcUsage
470                                    .with_span_context(span_context));
471                            }
472                            // Also, all members must have the same type.
473                            if members[0].ty != members[1].ty {
474                                return Err(VaryingError::BlendSrcOutputTypeMismatch {
475                                    blend_src_0_type: members[0].ty,
476                                    blend_src_1_type: members[1].ty,
477                                }
478                                .with_span_context(span_context));
479                            }
480                        }
481                    }
482                    _ => {
483                        if self.flags.contains(super::ValidationFlags::BINDINGS) {
484                            return Err(VaryingError::MissingBinding.with_span());
485                        }
486                    }
487                }
488                Ok(())
489            }
490        }
491    }
492}
493
494impl super::Validator {
495    pub(super) fn validate_global_var(
496        &self,
497        var: &crate::GlobalVariable,
498        gctx: crate::proc::GlobalCtx,
499        mod_info: &ModuleInfo,
500        global_expr_kind: &crate::proc::ExpressionKindTracker,
501    ) -> Result<(), GlobalVariableError> {
502        use super::TypeFlags;
503
504        log::debug!("var {var:?}");
505        let inner_ty = match gctx.types[var.ty].inner {
506            // A binding array is (mostly) supposed to behave the same as a
507            // series of individually bound resources, so we can (mostly)
508            // validate a `binding_array<T>` as if it were just a plain `T`.
509            crate::TypeInner::BindingArray { base, .. } => match var.space {
510                crate::AddressSpace::Storage { .. }
511                | crate::AddressSpace::Uniform
512                | crate::AddressSpace::Handle => base,
513                _ => return Err(GlobalVariableError::InvalidUsage(var.space)),
514            },
515            _ => var.ty,
516        };
517        let type_info = &self.types[inner_ty.index()];
518
519        let (required_type_flags, is_resource) = match var.space {
520            crate::AddressSpace::Function => {
521                return Err(GlobalVariableError::InvalidUsage(var.space))
522            }
523            crate::AddressSpace::Storage { access } => {
524                if let Err((ty_handle, disalignment)) = type_info.storage_layout {
525                    if self.flags.contains(super::ValidationFlags::STRUCT_LAYOUTS) {
526                        return Err(GlobalVariableError::Alignment(
527                            var.space,
528                            ty_handle,
529                            disalignment,
530                        ));
531                    }
532                }
533                if access == crate::StorageAccess::STORE {
534                    return Err(GlobalVariableError::StorageAddressSpaceWriteOnlyNotSupported);
535                }
536                (
537                    TypeFlags::DATA | TypeFlags::HOST_SHAREABLE | TypeFlags::CREATION_RESOLVED,
538                    true,
539                )
540            }
541            crate::AddressSpace::Uniform => {
542                if let Err((ty_handle, disalignment)) = type_info.uniform_layout {
543                    if self.flags.contains(super::ValidationFlags::STRUCT_LAYOUTS) {
544                        return Err(GlobalVariableError::Alignment(
545                            var.space,
546                            ty_handle,
547                            disalignment,
548                        ));
549                    }
550                }
551                (
552                    TypeFlags::DATA
553                        | TypeFlags::COPY
554                        | TypeFlags::SIZED
555                        | TypeFlags::HOST_SHAREABLE
556                        | TypeFlags::CREATION_RESOLVED,
557                    true,
558                )
559            }
560            crate::AddressSpace::Handle => {
561                match gctx.types[inner_ty].inner {
562                    crate::TypeInner::Image { class, .. } => match class {
563                        crate::ImageClass::Storage {
564                            format:
565                                crate::StorageFormat::R16Unorm
566                                | crate::StorageFormat::R16Snorm
567                                | crate::StorageFormat::Rg16Unorm
568                                | crate::StorageFormat::Rg16Snorm
569                                | crate::StorageFormat::Rgba16Unorm
570                                | crate::StorageFormat::Rgba16Snorm,
571                            ..
572                        } => {
573                            if !self
574                                .capabilities
575                                .contains(Capabilities::STORAGE_TEXTURE_16BIT_NORM_FORMATS)
576                            {
577                                return Err(GlobalVariableError::UnsupportedCapability(
578                                    Capabilities::STORAGE_TEXTURE_16BIT_NORM_FORMATS,
579                                ));
580                            }
581                        }
582                        _ => {}
583                    },
584                    crate::TypeInner::Sampler { .. }
585                    | crate::TypeInner::AccelerationStructure { .. }
586                    | crate::TypeInner::RayQuery { .. } => {}
587                    _ => {
588                        return Err(GlobalVariableError::InvalidType(var.space));
589                    }
590                }
591
592                (TypeFlags::empty(), true)
593            }
594            crate::AddressSpace::Private => (
595                TypeFlags::CONSTRUCTIBLE | TypeFlags::CREATION_RESOLVED,
596                false,
597            ),
598            crate::AddressSpace::WorkGroup => (TypeFlags::DATA | TypeFlags::SIZED, false),
599            crate::AddressSpace::PushConstant => {
600                if !self.capabilities.contains(Capabilities::PUSH_CONSTANT) {
601                    return Err(GlobalVariableError::UnsupportedCapability(
602                        Capabilities::PUSH_CONSTANT,
603                    ));
604                }
605                if let Err(ref err) = type_info.push_constant_compatibility {
606                    return Err(GlobalVariableError::InvalidPushConstantType(err.clone()));
607                }
608                (
609                    TypeFlags::DATA
610                        | TypeFlags::COPY
611                        | TypeFlags::HOST_SHAREABLE
612                        | TypeFlags::SIZED,
613                    false,
614                )
615            }
616        };
617
618        if !type_info.flags.contains(required_type_flags) {
619            return Err(GlobalVariableError::MissingTypeFlags {
620                seen: type_info.flags,
621                required: required_type_flags,
622            });
623        }
624
625        if is_resource != var.binding.is_some() {
626            if self.flags.contains(super::ValidationFlags::BINDINGS) {
627                return Err(GlobalVariableError::InvalidBinding);
628            }
629        }
630
631        if let Some(init) = var.init {
632            match var.space {
633                crate::AddressSpace::Private | crate::AddressSpace::Function => {}
634                _ => {
635                    return Err(GlobalVariableError::InitializerNotAllowed(var.space));
636                }
637            }
638
639            if !global_expr_kind.is_const_or_override(init) {
640                return Err(GlobalVariableError::InitializerExprType);
641            }
642
643            if !gctx.compare_types(
644                &crate::proc::TypeResolution::Handle(var.ty),
645                &mod_info[init],
646            ) {
647                return Err(GlobalVariableError::InitializerType);
648            }
649        }
650
651        Ok(())
652    }
653
654    pub(super) fn validate_entry_point(
655        &mut self,
656        ep: &crate::EntryPoint,
657        module: &crate::Module,
658        mod_info: &ModuleInfo,
659    ) -> Result<FunctionInfo, WithSpan<EntryPointError>> {
660        if ep.early_depth_test.is_some() {
661            let required = Capabilities::EARLY_DEPTH_TEST;
662            if !self.capabilities.contains(required) {
663                return Err(
664                    EntryPointError::Result(VaryingError::UnsupportedCapability(required))
665                        .with_span(),
666                );
667            }
668
669            if ep.stage != crate::ShaderStage::Fragment {
670                return Err(EntryPointError::UnexpectedEarlyDepthTest.with_span());
671            }
672        }
673
674        if ep.stage == crate::ShaderStage::Compute {
675            if ep
676                .workgroup_size
677                .iter()
678                .any(|&s| s == 0 || s > MAX_WORKGROUP_SIZE)
679            {
680                return Err(EntryPointError::OutOfRangeWorkgroupSize.with_span());
681            }
682        } else if ep.workgroup_size != [0; 3] {
683            return Err(EntryPointError::UnexpectedWorkgroupSize.with_span());
684        }
685
686        let mut info = self
687            .validate_function(&ep.function, module, mod_info, true)
688            .map_err(WithSpan::into_other)?;
689
690        {
691            use super::ShaderStages;
692
693            let stage_bit = match ep.stage {
694                crate::ShaderStage::Vertex => ShaderStages::VERTEX,
695                crate::ShaderStage::Fragment => ShaderStages::FRAGMENT,
696                crate::ShaderStage::Compute => ShaderStages::COMPUTE,
697                crate::ShaderStage::Task | crate::ShaderStage::Mesh => unreachable!(),
698            };
699
700            if !info.available_stages.contains(stage_bit) {
701                return Err(EntryPointError::ForbiddenStageOperations.with_span());
702            }
703        }
704
705        self.location_mask.clear();
706        let mut argument_built_ins = crate::FastHashSet::default();
707        // TODO: add span info to function arguments
708        for (index, fa) in ep.function.arguments.iter().enumerate() {
709            let mut ctx = VaryingContext {
710                stage: ep.stage,
711                output: false,
712                types: &module.types,
713                type_info: &self.types,
714                location_mask: &mut self.location_mask,
715                blend_src_mask: &mut self.blend_src_mask,
716                built_ins: &mut argument_built_ins,
717                capabilities: self.capabilities,
718                flags: self.flags,
719            };
720            ctx.validate(ep, fa.ty, fa.binding.as_ref())
721                .map_err_inner(|e| EntryPointError::Argument(index as u32, e).with_span())?;
722        }
723
724        self.location_mask.clear();
725        if let Some(ref fr) = ep.function.result {
726            let mut result_built_ins = crate::FastHashSet::default();
727            let mut ctx = VaryingContext {
728                stage: ep.stage,
729                output: true,
730                types: &module.types,
731                type_info: &self.types,
732                location_mask: &mut self.location_mask,
733                blend_src_mask: &mut self.blend_src_mask,
734                built_ins: &mut result_built_ins,
735                capabilities: self.capabilities,
736                flags: self.flags,
737            };
738            ctx.validate(ep, fr.ty, fr.binding.as_ref())
739                .map_err_inner(|e| EntryPointError::Result(e).with_span())?;
740            if ep.stage == crate::ShaderStage::Vertex
741                && !result_built_ins.contains(&crate::BuiltIn::Position { invariant: false })
742            {
743                return Err(EntryPointError::MissingVertexOutputPosition.with_span());
744            }
745            if !self.blend_src_mask.is_empty() {
746                info.dual_source_blending = true;
747            }
748        } else if ep.stage == crate::ShaderStage::Vertex {
749            return Err(EntryPointError::MissingVertexOutputPosition.with_span());
750        }
751
752        {
753            let mut used_push_constants = module
754                .global_variables
755                .iter()
756                .filter(|&(_, var)| var.space == crate::AddressSpace::PushConstant)
757                .map(|(handle, _)| handle)
758                .filter(|&handle| !info[handle].is_empty());
759            // Check if there is more than one push constant, and error if so.
760            // Use a loop for when returning multiple errors is supported.
761            if let Some(handle) = used_push_constants.nth(1) {
762                return Err(EntryPointError::MoreThanOnePushConstantUsed
763                    .with_span_handle(handle, &module.global_variables));
764            }
765        }
766
767        self.ep_resource_bindings.clear();
768        for (var_handle, var) in module.global_variables.iter() {
769            let usage = info[var_handle];
770            if usage.is_empty() {
771                continue;
772            }
773
774            let allowed_usage = match var.space {
775                crate::AddressSpace::Function => unreachable!(),
776                crate::AddressSpace::Uniform => GlobalUse::READ | GlobalUse::QUERY,
777                crate::AddressSpace::Storage { access } => storage_usage(access),
778                crate::AddressSpace::Handle => match module.types[var.ty].inner {
779                    crate::TypeInner::BindingArray { base, .. } => match module.types[base].inner {
780                        crate::TypeInner::Image {
781                            class: crate::ImageClass::Storage { access, .. },
782                            ..
783                        } => storage_usage(access),
784                        _ => GlobalUse::READ | GlobalUse::QUERY,
785                    },
786                    crate::TypeInner::Image {
787                        class: crate::ImageClass::Storage { access, .. },
788                        ..
789                    } => storage_usage(access),
790                    _ => GlobalUse::READ | GlobalUse::QUERY,
791                },
792                crate::AddressSpace::Private | crate::AddressSpace::WorkGroup => {
793                    GlobalUse::READ | GlobalUse::WRITE | GlobalUse::QUERY
794                }
795                crate::AddressSpace::PushConstant => GlobalUse::READ,
796            };
797            if !allowed_usage.contains(usage) {
798                log::warn!("\tUsage error for: {var:?}");
799                log::warn!("\tAllowed usage: {allowed_usage:?}, requested: {usage:?}");
800                return Err(EntryPointError::InvalidGlobalUsage(var_handle, usage)
801                    .with_span_handle(var_handle, &module.global_variables));
802            }
803
804            if let Some(ref bind) = var.binding {
805                if !self.ep_resource_bindings.insert(*bind) {
806                    if self.flags.contains(super::ValidationFlags::BINDINGS) {
807                        return Err(EntryPointError::BindingCollision(var_handle)
808                            .with_span_handle(var_handle, &module.global_variables));
809                    }
810                }
811            }
812        }
813
814        Ok(info)
815    }
816}
naga/valid/interface.rs

naga/valid/
interface.rs
