naga/valid/
interface.rs

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

naga/valid/
interface.rs
