mirror of
https://github.com/gfx-rs/wgpu.git
synced 2026-04-22 03:02:01 -04:00
Start wgsl-out (#727)
This commit is contained in:
@@ -39,6 +39,7 @@ deserialize = ["serde"]
|
||||
spv-in = ["petgraph", "spirv"]
|
||||
spv-out = ["spirv"]
|
||||
wgsl-in = ["codespan-reporting"]
|
||||
wgsl-out = []
|
||||
hlsl-out = []
|
||||
|
||||
[[bin]]
|
||||
|
||||
@@ -21,7 +21,7 @@ Rust | | | |
|
||||
Back-end | Status | Feature | Notes |
|
||||
--------------- | ------------------ | -------- | ----- |
|
||||
SPIR-V | :white_check_mark: | spv-out | |
|
||||
WGSL | | | |
|
||||
WGSL | :construction: | wgsl-out | |
|
||||
Metal | :white_check_mark: | msl-out | |
|
||||
HLSL | :construction: | hlsl-out | |
|
||||
GLSL | :ok: | glsl-out | |
|
||||
|
||||
@@ -268,6 +268,13 @@ fn main() {
|
||||
let hlsl = hlsl::write_string(&module).unwrap_pretty();
|
||||
fs::write(output_path, hlsl).unwrap();
|
||||
}
|
||||
#[cfg(feature = "wgsl-out")]
|
||||
"wgsl" => {
|
||||
use naga::back::wgsl;
|
||||
|
||||
let wgsl = wgsl::write_string(&module).unwrap_pretty();
|
||||
fs::write(output_path, wgsl).unwrap();
|
||||
}
|
||||
other => {
|
||||
let _ = params;
|
||||
panic!(
|
||||
|
||||
@@ -10,6 +10,8 @@ pub mod hlsl;
|
||||
pub mod msl;
|
||||
#[cfg(feature = "spv-out")]
|
||||
pub mod spv;
|
||||
#[cfg(feature = "wgsl-out")]
|
||||
pub mod wgsl;
|
||||
|
||||
impl crate::Expression {
|
||||
/// Returns the ref count, upon reaching which this expression
|
||||
|
||||
126
src/back/wgsl/keywords.rs
Normal file
126
src/back/wgsl/keywords.rs
Normal file
@@ -0,0 +1,126 @@
|
||||
// https://gpuweb.github.io/gpuweb/wgsl/#keyword-summary
|
||||
pub const RESERVED: &[&str] = &[
|
||||
// Type-defining keywords
|
||||
"ARRAY",
|
||||
"BOOL",
|
||||
"FLOAT32",
|
||||
"INT32",
|
||||
"MAT2x2",
|
||||
"MAT2x3",
|
||||
"MAT2x4",
|
||||
"MAT3x2",
|
||||
"MAT3x3",
|
||||
"MAT3x4",
|
||||
"MAT4x2",
|
||||
"MAT4x3",
|
||||
"MAT4x4",
|
||||
"POINTER",
|
||||
"SAMPLER",
|
||||
"SAMPLER_COMPARISON",
|
||||
"STRUCT",
|
||||
"TEXTURE_1D",
|
||||
"TEXTURE_2D",
|
||||
"TEXTURE_2D_ARRAY",
|
||||
"TEXTURE_3D",
|
||||
"TEXTURE_CUBE",
|
||||
"TEXTURE_CUBE_ARRAY",
|
||||
"TEXTURE_MULTISAMPLED_2D",
|
||||
"TEXTURE_STORAGE_1D",
|
||||
"TEXTURE_STORAGE_2D",
|
||||
"TEXTURE_STORAGE_2D_ARRAY",
|
||||
"TEXTURE_STORAGE_3D",
|
||||
"TEXTURE_DEPTH_2D",
|
||||
"TEXTURE_DEPTH_2D_ARRAY",
|
||||
"TEXTURE_DEPTH_CUBE",
|
||||
"TEXTURE_DEPTH_CUBE_ARRAY",
|
||||
"UINT32",
|
||||
"VEC2",
|
||||
"VEC3",
|
||||
"VEC4",
|
||||
// Other keywords
|
||||
"BITCAST",
|
||||
"BLOCK",
|
||||
"BREAK",
|
||||
"CASE",
|
||||
"CONTINUE",
|
||||
"CONTINUING",
|
||||
"DEFAULT",
|
||||
"DISCARD",
|
||||
"ELSE",
|
||||
"ELSE_IF",
|
||||
"ENABLE",
|
||||
"FALLTHROUGH",
|
||||
"FALSE",
|
||||
"FN",
|
||||
"FOR",
|
||||
"FUNCTION",
|
||||
"IF",
|
||||
"LET",
|
||||
"LOOP",
|
||||
"PRIVATE",
|
||||
"RETURN",
|
||||
"STORAGE",
|
||||
"SWITCH",
|
||||
"TRUE",
|
||||
"TYPE",
|
||||
"UNIFORM",
|
||||
"VAR",
|
||||
"WORKGROUP",
|
||||
// Image format keywords
|
||||
"R8UNORM",
|
||||
"R8SNORM",
|
||||
"R8UINT",
|
||||
"R8SINT",
|
||||
"R16UINT",
|
||||
"R16SINT",
|
||||
"R16FLOAT",
|
||||
"RG8UNORM",
|
||||
"RG8SNORM",
|
||||
"RG8UINT",
|
||||
"RG8SINT",
|
||||
"R32UINT",
|
||||
"R32SINT",
|
||||
"R32FLOAT",
|
||||
"RG16UINT",
|
||||
"RG16SINT",
|
||||
"RG16FLOAT",
|
||||
"RGBA8UNORM",
|
||||
"RGBA8UNORM-SRGB",
|
||||
"RGBA8SNORM",
|
||||
"RGBA8UINT",
|
||||
"RGBA8SINT",
|
||||
"BGRA8UNORM",
|
||||
"BGRA8UNORM-SRGB",
|
||||
"RGB10A2UNORM",
|
||||
"RG11B10FLOAT",
|
||||
"RG32UINT",
|
||||
"RG32SINT",
|
||||
"RG32FLOAT",
|
||||
"RGBA16UINT",
|
||||
"RGBA16SINT",
|
||||
"RGBA16FLOAT",
|
||||
"RGBA32UINT",
|
||||
"RGBA32SINT",
|
||||
"RGBA32FLOAT",
|
||||
// Reserved Keywords
|
||||
"asm",
|
||||
"bf16",
|
||||
"do",
|
||||
"enum",
|
||||
"f16",
|
||||
"f64",
|
||||
"i8",
|
||||
"i16",
|
||||
"i64",
|
||||
"const",
|
||||
"typedef",
|
||||
"u8",
|
||||
"u16",
|
||||
"u64",
|
||||
"unless",
|
||||
"using",
|
||||
"while",
|
||||
"regardless",
|
||||
"premerge",
|
||||
"handle",
|
||||
];
|
||||
23
src/back/wgsl/mod.rs
Normal file
23
src/back/wgsl/mod.rs
Normal file
@@ -0,0 +1,23 @@
|
||||
mod keywords;
|
||||
mod writer;
|
||||
|
||||
use std::io::Error as IoError;
|
||||
use std::string::FromUtf8Error;
|
||||
use thiserror::Error;
|
||||
|
||||
pub use writer::Writer;
|
||||
|
||||
#[derive(Error, Debug)]
|
||||
pub enum Error {
|
||||
#[error(transparent)]
|
||||
IoError(#[from] IoError),
|
||||
#[error(transparent)]
|
||||
Utf8(#[from] FromUtf8Error),
|
||||
}
|
||||
|
||||
pub fn write_string(module: &crate::Module) -> Result<String, Error> {
|
||||
let mut w = Writer::new(Vec::new());
|
||||
w.write(module)?;
|
||||
let output = String::from_utf8(w.finish())?;
|
||||
Ok(output)
|
||||
}
|
||||
144
src/back/wgsl/writer.rs
Normal file
144
src/back/wgsl/writer.rs
Normal file
@@ -0,0 +1,144 @@
|
||||
use super::Error;
|
||||
use crate::FastHashMap;
|
||||
use crate::{
|
||||
back::wgsl::keywords::RESERVED, Function, Module, ShaderStage, StructLevel, TypeInner,
|
||||
};
|
||||
use crate::{
|
||||
proc::{NameKey, Namer},
|
||||
StructMember,
|
||||
};
|
||||
use std::io::Write;
|
||||
|
||||
const _INDENT: &str = " ";
|
||||
|
||||
/// Shorthand result used internally by the backend
|
||||
type BackendResult = Result<(), Error>;
|
||||
|
||||
enum Decoration {
|
||||
VertexStage,
|
||||
FragmentStage,
|
||||
ComputeStage { workgroup_size: [u32; 3] },
|
||||
Block,
|
||||
}
|
||||
|
||||
pub struct Writer<W> {
|
||||
out: W,
|
||||
names: FastHashMap<NameKey, String>,
|
||||
namer: Namer,
|
||||
}
|
||||
|
||||
impl<W: Write> Writer<W> {
|
||||
pub fn new(out: W) -> Self {
|
||||
Writer {
|
||||
out,
|
||||
names: FastHashMap::default(),
|
||||
namer: Namer::default(),
|
||||
}
|
||||
}
|
||||
|
||||
pub fn write(&mut self, module: &Module) -> BackendResult {
|
||||
self.names.clear();
|
||||
self.namer.reset(module, RESERVED, &mut self.names);
|
||||
|
||||
// Write all structs
|
||||
for (handle, ty) in module.types.iter() {
|
||||
if let TypeInner::Struct {
|
||||
level, ref members, ..
|
||||
} = ty.inner
|
||||
{
|
||||
let name = &self.names[&NameKey::Type(handle)].clone();
|
||||
let block = level == StructLevel::Root;
|
||||
self.write_struct(name, block, members)?;
|
||||
writeln!(self.out)?;
|
||||
}
|
||||
}
|
||||
|
||||
for (_, ep) in module.entry_points.iter().enumerate() {
|
||||
let decoration = match ep.stage {
|
||||
ShaderStage::Vertex => Decoration::VertexStage,
|
||||
ShaderStage::Fragment => Decoration::FragmentStage,
|
||||
ShaderStage::Compute => Decoration::ComputeStage {
|
||||
workgroup_size: ep.workgroup_size,
|
||||
},
|
||||
};
|
||||
|
||||
self.write_decoration(decoration)?;
|
||||
// Add a newline after decoration
|
||||
writeln!(self.out)?;
|
||||
self.write_function(&ep.function)?;
|
||||
}
|
||||
|
||||
// Add a newline at the end of file
|
||||
writeln!(self.out)?;
|
||||
|
||||
Ok(())
|
||||
}
|
||||
|
||||
/// Helper method used to write structs
|
||||
/// https://gpuweb.github.io/gpuweb/wgsl/#functions
|
||||
///
|
||||
/// # Notes
|
||||
/// Ends in a newline
|
||||
fn write_function(&mut self, func: &Function) -> BackendResult {
|
||||
write!(self.out, "fn {}(", func.name.as_ref().unwrap())?; // TODO: unnamed function?
|
||||
write!(self.out, ")")?;
|
||||
|
||||
write!(self.out, "{{")?;
|
||||
write!(self.out, "}}")?;
|
||||
Ok(())
|
||||
}
|
||||
|
||||
/// Helper method to write a decoration
|
||||
///
|
||||
/// # Notes
|
||||
/// Adds no leading or trailing whitespace
|
||||
fn write_decoration(&mut self, decoration: Decoration) -> BackendResult {
|
||||
write!(self.out, "[[")?;
|
||||
match decoration {
|
||||
Decoration::VertexStage => write!(self.out, "stage(vertex)")?,
|
||||
Decoration::FragmentStage => write!(self.out, "stage(fragment)")?,
|
||||
Decoration::ComputeStage { workgroup_size } => {
|
||||
write!(
|
||||
self.out,
|
||||
"{}",
|
||||
format!(
|
||||
"stage(compute), workgroup_size({}, {}, {})",
|
||||
workgroup_size[0], workgroup_size[1], workgroup_size[2]
|
||||
)
|
||||
)?;
|
||||
}
|
||||
Decoration::Block => {
|
||||
write!(self.out, "block")?;
|
||||
}
|
||||
};
|
||||
write!(self.out, "]]")?;
|
||||
|
||||
Ok(())
|
||||
}
|
||||
|
||||
/// Helper method used to write structs
|
||||
///
|
||||
/// # Notes
|
||||
/// Ends in a newline
|
||||
fn write_struct(
|
||||
&mut self,
|
||||
name: &str,
|
||||
block: bool,
|
||||
_members: &[StructMember],
|
||||
) -> BackendResult {
|
||||
if block {
|
||||
self.write_decoration(Decoration::Block)?;
|
||||
writeln!(self.out)?;
|
||||
}
|
||||
write!(self.out, "struct {} {{", name)?;
|
||||
write!(self.out, "}}")?;
|
||||
|
||||
writeln!(self.out)?;
|
||||
|
||||
Ok(())
|
||||
}
|
||||
|
||||
pub fn finish(self) -> W {
|
||||
self.out
|
||||
}
|
||||
}
|
||||
Reference in New Issue
Block a user