Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Jump to bottom

Ray Tracing Support (glsl->spv) #2289

Draft
wants to merge 1 commit into
base: master
Choose a base branch
from
Draft

Ray Tracing Support (glsl->spv) #2289

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
9 changes: 8 additions & 1 deletion cli/src/bin/naga.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -287,7 +287,8 @@ fn run() -> Result<(), Box<dyn std::error::Error>> {
}
}
}
ext @ ("vert" | "frag" | "comp") => {
ext @ ("vert" | "frag" | "comp" | "rgen" | "rmiss" | "rcall" | "rchit" | "rahit"
| "rint") => {
let input = String::from_utf8(input)?;
let mut parser = naga::front::glsl::Frontend::default();

Expand All @@ -299,6 +300,12 @@ fn run() -> Result<(), Box<dyn std::error::Error>> {
"vert" => naga::ShaderStage::Vertex,
"frag" => naga::ShaderStage::Fragment,
"comp" => naga::ShaderStage::Compute,
"rgen" => naga::ShaderStage::RayGen,
"rmiss" => naga::ShaderStage::Miss,
"rcall" => naga::ShaderStage::Callable,
"rchit" => naga::ShaderStage::ClosestHit,
"rahit" => naga::ShaderStage::AnyHit,
"rint" => naga::ShaderStage::Intersection,
_ => unreachable!(),
},
defines: Default::default(),
Expand Down
21 changes: 21 additions & 0 deletions src/back/glsl/mod.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -107,6 +107,7 @@ impl crate::AddressSpace {
| crate::AddressSpace::Storage { .. }
| crate::AddressSpace::Handle
| crate::AddressSpace::PushConstant => false,
crate::AddressSpace::IncomingRayPayload => unimplemented!(),
}
}
}
Expand Down Expand Up @@ -335,6 +336,12 @@ impl fmt::Display for VaryingName<'_> {
crate::Binding::Location { location, .. } => {
let prefix = match (self.stage, self.output) {
(ShaderStage::Compute, _) => unreachable!(),
(ShaderStage::RayGen, _) => unreachable!(),
(ShaderStage::Miss, _) => unreachable!(),
(ShaderStage::Callable, _) => unreachable!(),
(ShaderStage::ClosestHit, _) => unreachable!(),
(ShaderStage::AnyHit, _) => unreachable!(),
(ShaderStage::Intersection, _) => unreachable!(),
// pipeline to vertex
(ShaderStage::Vertex, false) => "p2vs",
// vertex to fragment
Expand All @@ -361,6 +368,12 @@ impl ShaderStage {
ShaderStage::Compute => "cs",
ShaderStage::Fragment => "fs",
ShaderStage::Vertex => "vs",
ShaderStage::RayGen
| ShaderStage::Miss
| ShaderStage::Callable
| ShaderStage::ClosestHit
| ShaderStage::AnyHit
| ShaderStage::Intersection => unimplemented!(),
}
}
}
Expand Down Expand Up @@ -1030,6 +1043,7 @@ impl<'a, W: Write> Writer<'a, W> {
crate::AddressSpace::Function => unreachable!(),
// Textures and samplers are handled directly in `Writer::write`.
crate::AddressSpace::Handle => unreachable!(),
crate::AddressSpace::IncomingRayPayload => unimplemented!(),
}

Ok(())
Expand Down Expand Up @@ -1302,6 +1316,12 @@ impl<'a, W: Write> Writer<'a, W> {
ShaderStage::Vertex => output,
ShaderStage::Fragment => !output,
ShaderStage::Compute => false,
ShaderStage::RayGen
| ShaderStage::Miss
| ShaderStage::Callable
| ShaderStage::ClosestHit
| ShaderStage::AnyHit
| ShaderStage::Intersection => unimplemented!(),
};

// Write the I/O locations, if allowed
Expand Down Expand Up @@ -4010,6 +4030,7 @@ const fn glsl_storage_qualifier(space: crate::AddressSpace) -> Option<&'static s
As::Handle => Some("uniform"),
As::WorkGroup => Some("shared"),
As::PushConstant => Some("uniform"),
As::IncomingRayPayload => unimplemented!(),
}
}

Expand Down
6 changes: 6 additions & 0 deletions src/back/hlsl/mod.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -149,6 +149,12 @@ impl crate::ShaderStage {
Self::Vertex => "vs",
Self::Fragment => "ps",
Self::Compute => "cs",
Self::RayGen
| Self::Miss
| Self::Callable
| Self::ClosestHit
| Self::AnyHit
| Self::Intersection => unimplemented!(),
}
}
}
Expand Down
1 change: 1 addition & 0 deletions src/back/hlsl/writer.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -676,6 +676,7 @@ impl<'a, W: fmt::Write> super::Writer<'a, W> {
write!(self.out, "ConstantBuffer<")?;
"b"
}
crate::AddressSpace::IncomingRayPayload => unimplemented!(),
};

// If the global is a push constant write the type now because it will be a
Expand Down
10 changes: 10 additions & 0 deletions src/back/msl/writer.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -438,6 +438,7 @@ impl crate::AddressSpace {
| Self::PushConstant
| Self::Handle => true,
Self::Function => false,
Self::IncomingRayPayload => unimplemented!(),
}
}

Expand All @@ -455,6 +456,7 @@ impl crate::AddressSpace {
Self::Uniform | Self::PushConstant => false,
// Not applicable.
Self::Handle | Self::Function => false,
Self::IncomingRayPayload => unimplemented!(),
}
}

Expand All @@ -465,6 +467,7 @@ impl crate::AddressSpace {
Self::Storage { .. } => Some("device"),
Self::Private | Self::Function => Some("thread"),
Self::WorkGroup => Some("threadgroup"),
Self::IncomingRayPayload => unimplemented!(),
}
}
}
Expand Down Expand Up @@ -3432,6 +3435,7 @@ impl<W: Write> Writer<W> {
crate::AddressSpace::Function
| crate::AddressSpace::Private
| crate::AddressSpace::WorkGroup => {}
crate::AddressSpace::IncomingRayPayload => unimplemented!(),
}
}
if supports_array_length {
Expand Down Expand Up @@ -3464,6 +3468,12 @@ impl<W: Write> Writer<W> {
crate::ShaderStage::Compute { .. } => {
("kernel", LocationMode::Uniform, LocationMode::Uniform)
}
crate::ShaderStage::RayGen
| crate::ShaderStage::Miss
| crate::ShaderStage::Callable
| crate::ShaderStage::ClosestHit
| crate::ShaderStage::AnyHit
| crate::ShaderStage::Intersection => unimplemented!(),
};

// List all the Naga `EntryPoint`'s `Function`'s arguments,
Expand Down
1 change: 1 addition & 0 deletions src/back/spv/helpers.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -29,6 +29,7 @@ pub(super) const fn map_storage_class(space: crate::AddressSpace) -> spirv::Stor
crate::AddressSpace::Uniform => spirv::StorageClass::Uniform,
crate::AddressSpace::WorkGroup => spirv::StorageClass::Workgroup,
crate::AddressSpace::PushConstant => spirv::StorageClass::PushConstant,
crate::AddressSpace::IncomingRayPayload => spirv::StorageClass::IncomingRayPayloadKHR,
}
}

Expand Down
25 changes: 25 additions & 0 deletions src/back/spv/writer.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -766,6 +766,13 @@ impl Writer {
.to_words(&mut self.logical_layout.execution_modes);
spirv::ExecutionModel::GLCompute
}
// TODO
crate::ShaderStage::RayGen => spirv::ExecutionModel::RayGenerationKHR,
crate::ShaderStage::Miss => spirv::ExecutionModel::MissKHR,
crate::ShaderStage::Callable => spirv::ExecutionModel::CallableKHR,
crate::ShaderStage::ClosestHit => spirv::ExecutionModel::ClosestHitKHR,
crate::ShaderStage::AnyHit => spirv::ExecutionModel::AnyHitKHR,
crate::ShaderStage::Intersection => spirv::ExecutionModel::IntersectionKHR,
};
//self.check(exec_model.required_capabilities())?;

Expand Down Expand Up @@ -1587,6 +1594,24 @@ impl Writer {
}
};

match global_variable.space {
crate::AddressSpace::Function
| crate::AddressSpace::Private
| crate::AddressSpace::WorkGroup
| crate::AddressSpace::Uniform
| crate::AddressSpace::Storage { .. }
| crate::AddressSpace::Handle
| crate::AddressSpace::PushConstant => {}
crate::AddressSpace::IncomingRayPayload => {
self.require_any("Incoming Ray Payload", &[spirv::Capability::RayTracingKHR])?;
self.use_extension("SPV_KHR_ray_tracing");
}
};

if let Some(location) = global_variable.location {
self.decorate(id, Decoration::Location, &[location]);
};

let init_word = global_variable
.init
.map(|constant| self.constant_ids[constant.index()]);
Expand Down
19 changes: 19 additions & 0 deletions src/back/wgsl/writer.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -157,6 +157,12 @@ impl<W: Write> Writer<W> {
Attribute::Stage(ShaderStage::Compute),
Attribute::WorkGroupSize(ep.workgroup_size),
],
ShaderStage::RayGen
| ShaderStage::Miss
| ShaderStage::Callable
| ShaderStage::ClosestHit
| ShaderStage::AnyHit
| ShaderStage::Intersection => unimplemented!(),
};

self.write_attributes(&attributes)?;
Expand Down Expand Up @@ -209,6 +215,12 @@ impl<W: Write> Writer<W> {
ShaderStage::Compute => "ComputeOutput",
ShaderStage::Fragment => "FragmentOutput",
ShaderStage::Vertex => "VertexOutput",
ShaderStage::RayGen
| ShaderStage::Miss
| ShaderStage::Callable
| ShaderStage::ClosestHit
| ShaderStage::AnyHit
| ShaderStage::Intersection => unimplemented!(),
};

write!(self.out, "{name}")?;
Expand Down Expand Up @@ -343,6 +355,12 @@ impl<W: Write> Writer<W> {
ShaderStage::Vertex => "vertex",
ShaderStage::Fragment => "fragment",
ShaderStage::Compute => "compute",
ShaderStage::RayGen
| ShaderStage::Miss
| ShaderStage::Callable
| ShaderStage::ClosestHit
| ShaderStage::AnyHit
| ShaderStage::Intersection => unimplemented!(),
};
write!(self.out, "@{stage_str} ")?;
}
Expand Down Expand Up @@ -1931,6 +1949,7 @@ const fn address_space_str(
As::WorkGroup => "workgroup",
As::Handle => return (None, None),
As::Function => "function",
As::IncomingRayPayload => unimplemented!(),
}),
None,
)
Expand Down
1 change: 1 addition & 0 deletions src/front/glsl/lex.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -108,6 +108,7 @@ impl<'a> Iterator for Lexer<'a> {
// types
"void" => TokenValue::Void,
"struct" => TokenValue::Struct,
"rayPayloadInEXT" => TokenValue::RayPayloadInEXT,
word => match parse_type(word) {
Some(t) => TokenValue::TypeName(t),
None => TokenValue::Identifier(String::from(word)),
Expand Down
9 changes: 7 additions & 2 deletions src/front/glsl/parser/types.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -156,7 +156,8 @@ impl<'source> ParsingContext<'source> {
| TokenValue::Buffer
| TokenValue::Restrict
| TokenValue::MemoryQualifier(_)
| TokenValue::Layout => true,
| TokenValue::Layout
| TokenValue::RayPayloadInEXT => true,
_ => false,
})
}
Expand Down Expand Up @@ -210,7 +211,8 @@ impl<'source> ParsingContext<'source> {
| TokenValue::Out
| TokenValue::Uniform
| TokenValue::Shared
| TokenValue::Buffer => {
| TokenValue::Buffer
| TokenValue::RayPayloadInEXT => {
let storage = match token.value {
TokenValue::Const => StorageQualifier::Const,
TokenValue::In => StorageQualifier::Input,
Expand All @@ -226,6 +228,9 @@ impl<'source> ParsingContext<'source> {
access: crate::StorageAccess::all(),
})
}
TokenValue::RayPayloadInEXT => {
StorageQualifier::AddressSpace(AddressSpace::IncomingRayPayload)
}
_ => unreachable!(),
};

Expand Down
1 change: 1 addition & 0 deletions src/front/glsl/token.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -36,6 +36,7 @@ pub enum TokenValue {
Buffer,
Const,
Shared,
RayPayloadInEXT,

Restrict,
/// A `glsl` memory qualifier such as `writeonly`
Expand Down
15 changes: 15 additions & 0 deletions src/front/glsl/variables.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -58,6 +58,7 @@ impl Frontend {
name: Some(name.into()),
space: AddressSpace::Private,
binding: None,
location: None,
ty,
init: None,
},
Expand Down Expand Up @@ -462,6 +463,7 @@ impl Frontend {
name: name.clone(),
space: AddressSpace::Private,
binding: None,
location: None,
ty,
init,
},
Expand Down Expand Up @@ -600,11 +602,24 @@ impl Frontend {
_ => None,
};

let location = match space {
AddressSpace::IncomingRayPayload => {
// TODO: glslang seems to use a counter for variables without
// explicit location (even if that causes collisions)
let location = qualifiers
.uint_layout_qualifier("location", &mut self.errors)
.unwrap_or(0);
Some(location)
}
_ => None,
};

let handle = self.module.global_variables.append(
GlobalVariable {
name: name.clone(),
space,
binding,
location,
ty,
init,
},
Expand Down
3 changes: 3 additions & 0 deletions src/front/spv/mod.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -5003,6 +5003,7 @@ impl<I: Iterator<Item = u32>> Frontend<I> {
}
let var = crate::GlobalVariable {
binding: dec.resource_binding(),
location: None,
name: dec.name,
space,
ty,
Expand Down Expand Up @@ -5048,6 +5049,7 @@ impl<I: Iterator<Item = u32>> Frontend<I> {
name: dec.name.clone(),
space: crate::AddressSpace::Private,
binding: None,
location: None,
ty,
init: None,
};
Expand Down Expand Up @@ -5121,6 +5123,7 @@ impl<I: Iterator<Item = u32>> Frontend<I> {
name: dec.name,
space: crate::AddressSpace::Private,
binding: None,
location: None,
ty,
init,
};
Expand Down
1 change: 1 addition & 0 deletions src/front/wgsl/lower/mod.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -666,6 +666,7 @@ impl<'source, 'temp> Lowerer<'source, 'temp> {
name: Some(v.name.name.to_string()),
space: v.space,
binding: v.binding.clone(),
location: None,
ty,
init,
},
Expand Down
10 changes: 10 additions & 0 deletions src/lib.rs
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -302,6 +302,12 @@ pub enum ShaderStage {
Vertex,
Fragment,
Compute,
RayGen,
Miss,
Callable,
ClosestHit,
AnyHit,
Intersection,
}

/// Addressing space of variables.
Expand All @@ -324,6 +330,8 @@ pub enum AddressSpace {
Handle,
/// Push constants.
PushConstant,
/// Incoming ray payload for SPV_KHR_ray_tracing
IncomingRayPayload,
}

/// Built-in inputs and outputs.
Expand Down Expand Up @@ -861,6 +869,8 @@ pub struct GlobalVariable {
pub space: AddressSpace,
/// For resources, defines the binding point.
pub binding: Option<ResourceBinding>,
/// For some ray-tracing storage qualifiers, define the location
pub location: Option<u32>,
Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

does this need to work differently from the existing mechanism of passing locations? Currently, struct Binding has a location, and it's carried for entry point inputs/outputs. The incoming ray payload should probably just be an argument to the entry point.

Copy link
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Moreover, we should probably consider just merging the IncompingRayPayload address space into Input, since it's non-ambiguous and likely has the same constraints.

Copy link
Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

All new storage classes are somewhere between StorageQualifier::Output and StorageQualifier::AddressSpace.
They are read and write, have a location, no interpolation, no binding.
Additionally storage classes are mapped from AddressSpaces.
So StorageQualifier::AddressSpace seemed to be the better fit.
A new StorageQualifier may even be best.

/// The type of this variable.
pub ty: Handle<Type>,
/// Initial value for this variable.
Expand Down
Loading