Skip to content

Commit

Permalink
Merge branch 'trunk' into idx-offset-to-first-idx
Browse files Browse the repository at this point in the history
  • Loading branch information
@Vecvec
Vecvec authored Jan 8, 2025
2 parents 67c4413 + a8a9173 commit 042d872
Show file tree
Hide file tree
Showing 30 changed files with 485 additions and 326 deletions.
2 changes: 1 addition & 1 deletion .github/workflows/ci.yml
View file
Original file line number Diff line number Diff line change
Expand Up @@ -48,7 +48,7 @@ env:
CARGO_INCREMENTAL: false
CARGO_TERM_COLOR: always
WGPU_DX12_COMPILER: dxc
RUST_LOG: info
RUST_LOG: debug
RUST_BACKTRACE: full
PKG_CONFIG_ALLOW_CROSS: 1 # allow android to work
RUSTFLAGS: -D warnings
Expand Down
3 changes: 1 addition & 2 deletions Cargo.lock
View file

Some generated files are not rendered by default. Learn more about how customized files appear on GitHub.

2 changes: 1 addition & 1 deletion Cargo.toml
View file
Original file line number Diff line number Diff line change
Expand Up @@ -142,9 +142,9 @@ wgpu-types = { version = "23.0.0", path = "./wgpu-types" }
winit = { version = "0.29", features = ["android-native-activity"] }

# Metal dependencies
metal = { version = "0.30.0", git = "https://github.com/gfx-rs/metal-rs.git", rev = "ef768ff9d7" }
block = "0.1"
core-graphics-types = "0.1"
metal = { version = "0.30.0" }
objc = "0.2.5"

# Vulkan dependencies
Expand Down
9 changes: 9 additions & 0 deletions benches/benches/bind_groups.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -62,6 +62,15 @@ impl BindGroupState {
fn run_bench(ctx: &mut Criterion) {
let state = Lazy::new(BindGroupState::new);

if !state
.device_state
.device
.features()
.contains(wgpu::Features::TEXTURE_BINDING_ARRAY)
{
return;
}

let mut group = ctx.benchmark_group("Bind Group Creation");

for count in [5, 50, 500, 5_000, 50_000] {
Expand Down
13 changes: 8 additions & 5 deletions naga/src/back/glsl/mod.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -747,14 +747,17 @@ impl<'a, W: Write> Writer<'a, W> {
// Write functions to create special types.
for (type_key, struct_ty) in self.module.special_types.predeclared_types.iter() {
match type_key {
&crate::PredeclaredType::ModfResult { size, width }
| &crate::PredeclaredType::FrexpResult { size, width } => {
&crate::PredeclaredType::ModfResult { size, scalar }
| &crate::PredeclaredType::FrexpResult { size, scalar } => {
let arg_type_name_owner;
let arg_type_name = if let Some(size) = size {
arg_type_name_owner =
format!("{}vec{}", if width == 8 { "d" } else { "" }, size as u8);
arg_type_name_owner = format!(
"{}vec{}",
if scalar.width == 8 { "d" } else { "" },
size as u8
);
&arg_type_name_owner
} else if width == 8 {
} else if scalar.width == 8 {
"double"
} else {
"float"
Expand Down
8 changes: 4 additions & 4 deletions naga/src/back/hlsl/help.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -796,17 +796,17 @@ impl<W: Write> super::Writer<'_, W> {
pub(super) fn write_special_functions(&mut self, module: &crate::Module) -> BackendResult {
for (type_key, struct_ty) in module.special_types.predeclared_types.iter() {
match type_key {
&crate::PredeclaredType::ModfResult { size, width }
| &crate::PredeclaredType::FrexpResult { size, width } => {
&crate::PredeclaredType::ModfResult { size, scalar }
| &crate::PredeclaredType::FrexpResult { size, scalar } => {
let arg_type_name_owner;
let arg_type_name = if let Some(size) = size {
arg_type_name_owner = format!(
"{}{}",
if width == 8 { "double" } else { "float" },
if scalar.width == 8 { "double" } else { "float" },
size as u8
);
&arg_type_name_owner
} else if width == 8 {
} else if scalar.width == 8 {
"double"
} else {
"float"
Expand Down
1 change: 1 addition & 0 deletions naga/src/back/msl/keywords.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -341,4 +341,5 @@ pub const RESERVED: &[&str] = &[
"DefaultConstructible",
super::writer::FREXP_FUNCTION,
super::writer::MODF_FUNCTION,
super::writer::ARGUMENT_BUFFER_WRAPPER_STRUCT,
];
2 changes: 0 additions & 2 deletions naga/src/back/msl/mod.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -59,8 +59,6 @@ pub struct BindTarget {
pub buffer: Option<Slot>,
pub texture: Option<Slot>,
pub sampler: Option<BindSamplerTarget>,
/// If the binding is an unsized binding array, this overrides the size.
pub binding_array_size: Option<u32>,
pub mutable: bool,
}

Expand Down
61 changes: 38 additions & 23 deletions naga/src/back/msl/writer.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -36,6 +36,14 @@ const RAY_QUERY_FUN_MAP_INTERSECTION: &str = "_map_intersection_type";
pub(crate) const ATOMIC_COMP_EXCH_FUNCTION: &str = "naga_atomic_compare_exchange_weak_explicit";
pub(crate) const MODF_FUNCTION: &str = "naga_modf";
pub(crate) const FREXP_FUNCTION: &str = "naga_frexp";
/// For some reason, Metal does not let you have `metal::texture<..>*` as a buffer argument.
/// However, if you put that texture inside a struct, everything is totally fine. This
/// baffles me to no end.
///
/// As such, we wrap all argument buffers in a struct that has a single generic `<T>` field.
/// This allows `NagaArgumentBufferWrapper<metal::texture<..>>*` to work. The astute among
/// you have noticed that this should be exactly the same to the compiler, and you're correct.
pub(crate) const ARGUMENT_BUFFER_WRAPPER_STRUCT: &str = "NagaArgumentBufferWrapper";

/// Write the Metal name for a Naga numeric type: scalar, vector, or matrix.
///
Expand Down Expand Up @@ -275,24 +283,17 @@ impl Display for TypeContext<'_> {
crate::TypeInner::RayQuery => {
write!(out, "{RAY_QUERY_TYPE}")
}
crate::TypeInner::BindingArray { base, size } => {
crate::TypeInner::BindingArray { base, .. } => {
let base_tyname = Self {
handle: base,
first_time: false,
..*self
};

if let Some(&super::ResolvedBinding::Resource(super::BindTarget {
binding_array_size: Some(override_size),
..
})) = self.binding
{
write!(out, "{NAMESPACE}::array<{base_tyname}, {override_size}>")
} else if let crate::ArraySize::Constant(size) = size {
write!(out, "{NAMESPACE}::array<{base_tyname}, {size}>")
} else {
unreachable!("metal requires all arrays be constant sized");
}
write!(
out,
"constant {ARGUMENT_BUFFER_WRAPPER_STRUCT}<{base_tyname}>*"
)
}
}
}
Expand Down Expand Up @@ -2552,6 +2553,8 @@ impl<W: Write> Writer<W> {
} => true,
_ => false,
};
let accessing_wrapped_binding_array =
matches!(*base_ty, crate::TypeInner::BindingArray { .. });

self.put_access_chain(base, policy, context)?;
if accessing_wrapped_array {
Expand Down Expand Up @@ -2588,6 +2591,10 @@ impl<W: Write> Writer<W> {

write!(self.out, "]")?;

if accessing_wrapped_binding_array {
write!(self.out, ".{WRAPPED_ARRAY_FIELD}")?;
}

Ok(())
}

Expand Down Expand Up @@ -3701,7 +3708,18 @@ impl<W: Write> Writer<W> {
}

fn write_type_defs(&mut self, module: &crate::Module) -> BackendResult {
let mut generated_argument_buffer_wrapper = false;
for (handle, ty) in module.types.iter() {
if let crate::TypeInner::BindingArray { .. } = ty.inner {
if !generated_argument_buffer_wrapper {
writeln!(self.out, "template <typename T>")?;
writeln!(self.out, "struct {ARGUMENT_BUFFER_WRAPPER_STRUCT} {{")?;
writeln!(self.out, "{}T {WRAPPED_ARRAY_FIELD};", back::INDENT)?;
writeln!(self.out, "}};")?;
generated_argument_buffer_wrapper = true;
}
}

if !ty.needs_alias() {
continue;
}
Expand Down Expand Up @@ -3830,17 +3848,17 @@ impl<W: Write> Writer<W> {
// Write functions to create special types.
for (type_key, struct_ty) in module.special_types.predeclared_types.iter() {
match type_key {
&crate::PredeclaredType::ModfResult { size, width }
| &crate::PredeclaredType::FrexpResult { size, width } => {
&crate::PredeclaredType::ModfResult { size, scalar }
| &crate::PredeclaredType::FrexpResult { size, scalar } => {
let arg_type_name_owner;
let arg_type_name = if let Some(size) = size {
arg_type_name_owner = format!(
"{NAMESPACE}::{}{}",
if width == 8 { "double" } else { "float" },
if scalar.width == 8 { "double" } else { "float" },
size as u8
);
&arg_type_name_owner
} else if width == 8 {
} else if scalar.width == 8 {
"double"
} else {
"float"
Expand Down Expand Up @@ -5132,13 +5150,10 @@ template <typename A>
let target = options.get_resource_binding_target(ep, br);
let good = match target {
Some(target) => {
let binding_ty = match module.types[var.ty].inner {
crate::TypeInner::BindingArray { base, .. } => {
&module.types[base].inner
}
ref ty => ty,
};
match *binding_ty {
// We intentionally don't dereference binding_arrays here,
// so that binding arrays fall to the buffer location.

match module.types[var.ty].inner {
crate::TypeInner::Image { .. } => target.texture.is_some(),
crate::TypeInner::Sampler { .. } => {
target.sampler.is_some()
Expand Down
34 changes: 14 additions & 20 deletions naga/src/front/type_gen.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -298,11 +298,11 @@ impl crate::Module {
},
}
}
crate::PredeclaredType::ModfResult { size, width } => {
crate::PredeclaredType::ModfResult { size, scalar } => {
let float_ty = self.types.insert(
crate::Type {
name: None,
inner: crate::TypeInner::Scalar(crate::Scalar::float(width)),
inner: crate::TypeInner::Scalar(scalar),
},
Span::UNDEFINED,
);
Expand All @@ -311,23 +311,20 @@ impl crate::Module {
let vec_ty = self.types.insert(
crate::Type {
name: None,
inner: crate::TypeInner::Vector {
size,
scalar: crate::Scalar::float(width),
},
inner: crate::TypeInner::Vector { size, scalar },
},
Span::UNDEFINED,
);
(vec_ty, size as u32 * width as u32)
(vec_ty, size as u32 * scalar.width as u32)
} else {
(float_ty, width as u32)
(float_ty, scalar.width as u32)
};

let mut type_name = "__modf_result_".to_string();
if let Some(size) = size {
let _ = write!(type_name, "vec{}_", size as u8);
}
let _ = write!(type_name, "f{}", width * 8);
let _ = write!(type_name, "f{}", scalar.width * 8);

crate::Type {
name: Some(type_name),
Expand All @@ -350,11 +347,11 @@ impl crate::Module {
},
}
}
crate::PredeclaredType::FrexpResult { size, width } => {
crate::PredeclaredType::FrexpResult { size, scalar } => {
let float_ty = self.types.insert(
crate::Type {
name: None,
inner: crate::TypeInner::Scalar(crate::Scalar::float(width)),
inner: crate::TypeInner::Scalar(scalar),
},
Span::UNDEFINED,
);
Expand All @@ -364,7 +361,7 @@ impl crate::Module {
name: None,
inner: crate::TypeInner::Scalar(crate::Scalar {
kind: crate::ScalarKind::Sint,
width,
width: scalar.width,
}),
},
Span::UNDEFINED,
Expand All @@ -374,10 +371,7 @@ impl crate::Module {
let vec_float_ty = self.types.insert(
crate::Type {
name: None,
inner: crate::TypeInner::Vector {
size,
scalar: crate::Scalar::float(width),
},
inner: crate::TypeInner::Vector { size, scalar },
},
Span::UNDEFINED,
);
Expand All @@ -388,22 +382,22 @@ impl crate::Module {
size,
scalar: crate::Scalar {
kind: crate::ScalarKind::Sint,
width,
width: scalar.width,
},
},
},
Span::UNDEFINED,
);
(vec_float_ty, vec_int_ty, size as u32 * width as u32)
(vec_float_ty, vec_int_ty, size as u32 * scalar.width as u32)
} else {
(float_ty, int_ty, width as u32)
(float_ty, int_ty, scalar.width as u32)
};

let mut type_name = "__frexp_result_".to_string();
if let Some(size) = size {
let _ = write!(type_name, "vec{}_", size as u8);
}
let _ = write!(type_name, "f{}", width * 8);
let _ = write!(type_name, "f{}", scalar.width * 8);

crate::Type {
name: Some(type_name),
Expand Down
20 changes: 7 additions & 13 deletions naga/src/front/wgsl/lower/mod.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -2060,11 +2060,9 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {

lowered_base.map(|base| crate::Expression::AccessIndex { base, index })
}
crate::TypeInner::Vector { .. } | crate::TypeInner::Matrix { .. } => {
crate::TypeInner::Vector { .. } => {
match Components::new(field.name, field.span)? {
Components::Swizzle { size, pattern } => {
// Swizzles aren't allowed on matrices, but
// validation will catch that.
Typed::Plain(crate::Expression::Swizzle {
size,
vector: ctx.apply_load_rule(lowered_base)?,
Expand Down Expand Up @@ -2304,22 +2302,18 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
args.finish()?;

if fun == crate::MathFunction::Modf || fun == crate::MathFunction::Frexp {
if let Some((size, width)) = match *resolve_inner!(ctx, arg) {
crate::TypeInner::Scalar(crate::Scalar { width, .. }) => {
Some((None, width))
if let Some((size, scalar)) = match *resolve_inner!(ctx, arg) {
crate::TypeInner::Scalar(scalar) => Some((None, scalar)),
crate::TypeInner::Vector { size, scalar, .. } => {
Some((Some(size), scalar))
}
crate::TypeInner::Vector {
size,
scalar: crate::Scalar { width, .. },
..
} => Some((Some(size), width)),
_ => None,
} {
ctx.module.generate_predeclared_type(
if fun == crate::MathFunction::Modf {
crate::PredeclaredType::ModfResult { size, width }
crate::PredeclaredType::ModfResult { size, scalar }
} else {
crate::PredeclaredType::FrexpResult { size, width }
crate::PredeclaredType::FrexpResult { size, scalar }
},
);
}
Expand Down
8 changes: 8 additions & 0 deletions naga/src/front/wgsl/tests.rs
View file
Original file line number Diff line number Diff line change
Expand Up @@ -399,6 +399,14 @@ fn parse_postfix() {
}",
)
.unwrap();

let err = parse_str(
"fn foo() {
let v = mat4x4<f32>().x;
}",
)
.unwrap_err();
assert_eq!(err.message(), "invalid field accessor `x`");
}

#[test]
Expand Down
Loading

0 comments on commit 042d872

Please sign in to comment.