1use blade_graphics as gpu;
2
3use crate::{CameraParams, ColorConfig, EmitterShape, ParticleEffect};
4
5#[repr(C)]
6#[derive(Clone, Copy, bytemuck::Zeroable, bytemuck::Pod)]
7struct EmitParams {
8 origin: [f32; 3],
9 emitter_radius: f32,
10 direction: [f32; 3],
11 cone_half_angle_cos: f32,
12 colors: [u32; 4],
13 color_count: u32,
14 emit_count: u32,
15 life_min: f32,
16 life_max: f32,
17 speed_min: f32,
18 speed_max: f32,
19 scale_min: f32,
20 scale_max: f32,
21}
22
23#[repr(C)]
24#[derive(Clone, Copy, bytemuck::Zeroable, bytemuck::Pod)]
25struct UpdateParams {
26 time_delta: f32,
27}
28
29#[derive(blade_macros::ShaderData)]
30struct MainData {
31 particles: gpu::BufferPiece,
32 free_list: gpu::BufferPiece,
33}
34
35#[derive(blade_macros::ShaderData)]
36struct EmitData {
37 emit_params: EmitParams,
38}
39
40#[derive(blade_macros::ShaderData)]
41struct UpdateData {
42 update_params: UpdateParams,
43}
44
45#[derive(blade_macros::ShaderData)]
46struct DrawData {
47 draw_particles: gpu::BufferPiece,
48 camera: CameraParams,
49}
50
51pub struct PipelineDesc<'a> {
52 pub name: &'a str,
53 pub draw_format: gpu::TextureFormat,
54 pub depth_format: Option<gpu::TextureFormat>,
55 pub sample_count: u32,
56}
57
58pub struct ParticlePipeline {
61 reset_pipeline: gpu::ComputePipeline,
62 emit_pipeline: gpu::ComputePipeline,
63 update_pipeline: gpu::ComputePipeline,
64 draw_pipeline: gpu::RenderPipeline,
65 particle_size: u32,
66}
67
68impl ParticlePipeline {
69 pub fn new(context: &gpu::Context, desc: PipelineDesc) -> Self {
70 let source = include_str!("particle.wgsl");
71 let shader = context.create_shader(gpu::ShaderDesc {
72 source,
73 naga_module: None,
74 });
75
76 let particle_size = shader.get_struct_size("Particle");
77
78 let main_layout = <MainData as gpu::ShaderData>::layout();
79 let emit_layout = <EmitData as gpu::ShaderData>::layout();
80 let update_layout = <UpdateData as gpu::ShaderData>::layout();
81 let draw_layout = <DrawData as gpu::ShaderData>::layout();
82
83 let reset_pipeline = context.create_compute_pipeline(gpu::ComputePipelineDesc {
84 name: &format!("{} - reset", desc.name),
85 data_layouts: &[&main_layout],
86 compute: shader.at("reset"),
87 });
88 let emit_pipeline = context.create_compute_pipeline(gpu::ComputePipelineDesc {
89 name: &format!("{} - emit", desc.name),
90 data_layouts: &[&main_layout, &emit_layout],
91 compute: shader.at("emit"),
92 });
93 let update_pipeline = context.create_compute_pipeline(gpu::ComputePipelineDesc {
94 name: &format!("{} - update", desc.name),
95 data_layouts: &[&main_layout, &update_layout],
96 compute: shader.at("update"),
97 });
98 let draw_pipeline = context.create_render_pipeline(gpu::RenderPipelineDesc {
99 name: &format!("{} - draw", desc.name),
100 data_layouts: &[&draw_layout],
101 primitive: gpu::PrimitiveState {
102 topology: gpu::PrimitiveTopology::TriangleStrip,
103 ..Default::default()
104 },
105 vertex: shader.at("draw_vs"),
106 vertex_fetches: &[],
107 fragment: Some(shader.at("draw_fs")),
108 color_targets: &[gpu::ColorTargetState {
109 format: desc.draw_format,
110 blend: Some(gpu::BlendState::ALPHA_BLENDING),
111 write_mask: gpu::ColorWrites::default(),
112 }],
113 depth_stencil: desc.depth_format.map(|format| gpu::DepthStencilState {
114 format,
115 depth_write_enabled: false,
116 depth_compare: gpu::CompareFunction::LessEqual,
117 stencil: gpu::StencilState::default(),
118 bias: gpu::DepthBiasState::default(),
119 }),
120 multisample_state: gpu::MultisampleState {
121 sample_count: desc.sample_count,
122 ..Default::default()
123 },
124 });
125
126 Self {
127 reset_pipeline,
128 emit_pipeline,
129 update_pipeline,
130 draw_pipeline,
131 particle_size,
132 }
133 }
134
135 pub fn destroy(&mut self, context: &gpu::Context) {
136 context.destroy_compute_pipeline(&mut self.reset_pipeline);
137 context.destroy_compute_pipeline(&mut self.emit_pipeline);
138 context.destroy_compute_pipeline(&mut self.update_pipeline);
139 context.destroy_render_pipeline(&mut self.draw_pipeline);
140 }
141
142 pub fn create_system(
144 &self,
145 context: &gpu::Context,
146 name: &str,
147 effect: &ParticleEffect,
148 ) -> ParticleSystem {
149 let wg_width = self.reset_pipeline.get_workgroup_size()[0] as usize;
150 let capacity = ((effect.capacity as usize - 1) | (wg_width - 1)) + 1;
151 let particle_buf = context.create_buffer(gpu::BufferDesc {
152 name,
153 size: capacity as u64 * self.particle_size as u64,
154 memory: gpu::Memory::Device,
155 });
156 let free_list_buf = context.create_buffer(gpu::BufferDesc {
157 name: &format!("{} - free list", name),
158 size: 4 + capacity as u64 * 4,
159 memory: gpu::Memory::Device,
160 });
161
162 ParticleSystem {
163 capacity,
164 effect: effect.clone(),
165 origin: [0.0; 3],
166 axis: [0.0, 1.0, 0.0],
167 particle_buf,
168 free_list_buf,
169 emit_accumulator: 0.0,
170 pending_bursts: Vec::new(),
171 needs_reset: true,
172 }
173 }
174}
175
176struct PendingBurst {
177 count: u32,
178 position: [f32; 3],
179}
180
181pub struct ParticleSystem {
184 capacity: usize,
185 pub effect: ParticleEffect,
186 pub origin: [f32; 3],
188 pub axis: [f32; 3],
192 particle_buf: gpu::Buffer,
193 free_list_buf: gpu::Buffer,
194 emit_accumulator: f32,
195 pending_bursts: Vec<PendingBurst>,
196 needs_reset: bool,
197}
198
199impl ParticleSystem {
200 pub fn destroy(&mut self, context: &gpu::Context) {
201 context.destroy_buffer(self.particle_buf);
202 context.destroy_buffer(self.free_list_buf);
203 }
204
205 fn main_data(&self) -> MainData {
206 MainData {
207 particles: self.particle_buf.into(),
208 free_list: self.free_list_buf.into(),
209 }
210 }
211
212 fn make_emit_params(&self, count: u32, position: [f32; 3]) -> EmitParams {
213 let emitter_radius = match self.effect.emitter.shape {
214 EmitterShape::Point => 0.0,
215 EmitterShape::Sphere { radius } => radius,
216 };
217
218 let (colors, color_count) = match self.effect.particle.color {
219 ColorConfig::Solid(c) => {
220 let packed = pack_color(c);
221 ([packed, packed, packed, packed], 1u32)
222 }
223 ColorConfig::Palette(ref palette) => {
224 let mut colors = [0u32; 4];
225 let count = palette.len().min(4);
226 for i in 0..count {
227 colors[i] = pack_color(palette[i]);
228 }
229 for i in count..4 {
230 colors[i] = colors[0];
231 }
232 (colors, count as u32)
233 }
234 };
235
236 EmitParams {
237 origin: position,
238 emitter_radius,
239 direction: self.axis,
240 cone_half_angle_cos: self.effect.emitter.cone_angle.cos(),
241 colors,
242 color_count,
243 emit_count: count,
244 life_min: self.effect.particle.life[0],
245 life_max: self.effect.particle.life[1],
246 speed_min: self.effect.particle.speed[0],
247 speed_max: self.effect.particle.speed[1],
248 scale_min: self.effect.particle.scale[0],
249 scale_max: self.effect.particle.scale[1],
250 }
251 }
252
253 pub fn burst(&mut self, count: u32, position: [f32; 3]) {
255 self.pending_bursts.push(PendingBurst { count, position });
256 }
257
258 pub fn update(
260 &mut self,
261 pipeline: &ParticlePipeline,
262 encoder: &mut gpu::CommandEncoder,
263 dt: f32,
264 ) {
265 if self.needs_reset {
266 let mut pass = encoder.compute("particle reset");
267 let mut pc = pass.with(&pipeline.reset_pipeline);
268 pc.bind(0, &self.main_data());
269 let group_size = pipeline.reset_pipeline.get_workgroup_size();
270 let group_count = (self.capacity as u32).div_ceil(group_size[0]);
271 pc.dispatch([group_count, 1, 1]);
272 self.needs_reset = false;
273 }
274
275 let main_data = self.main_data();
276
277 {
279 let mut pass = encoder.compute("particle update");
280 let mut pc = pass.with(&pipeline.update_pipeline);
281 pc.bind(0, &main_data);
282 pc.bind(
283 1,
284 &UpdateData {
285 update_params: UpdateParams { time_delta: dt },
286 },
287 );
288 let group_size = pipeline.update_pipeline.get_workgroup_size();
289 let group_count = self.capacity as u32 / group_size[0];
290 pc.dispatch([group_count, 1, 1]);
291 }
292
293 if self.effect.emitter.rate > 0.0 {
295 self.emit_accumulator += self.effect.emitter.rate * dt;
296 let emit_count = self.emit_accumulator as u32;
297 if emit_count > 0 {
298 self.emit_accumulator -= emit_count as f32;
299 let params = self.make_emit_params(emit_count, self.origin);
300 let wg_size = pipeline.emit_pipeline.get_workgroup_size()[0];
301 let groups = emit_count.div_ceil(wg_size);
302 let mut pass = encoder.compute("particle emit continuous");
303 let mut pc = pass.with(&pipeline.emit_pipeline);
304 pc.bind(0, &main_data);
305 pc.bind(
306 1,
307 &EmitData {
308 emit_params: params,
309 },
310 );
311 pc.dispatch([groups, 1, 1]);
312 }
313 }
314
315 let bursts: Vec<_> = self.pending_bursts.drain(..).collect();
317 for burst in bursts {
318 let params = self.make_emit_params(burst.count, burst.position);
319 let wg_size = pipeline.emit_pipeline.get_workgroup_size()[0];
320 let groups = burst.count.div_ceil(wg_size);
321 let mut pass = encoder.compute("particle emit burst");
322 let mut pc = pass.with(&pipeline.emit_pipeline);
323 pc.bind(0, &main_data);
324 pc.bind(
325 1,
326 &EmitData {
327 emit_params: params,
328 },
329 );
330 pc.dispatch([groups, 1, 1]);
331 }
332 }
333
334 pub fn draw(
336 &self,
337 pipeline: &ParticlePipeline,
338 pass: &mut gpu::RenderCommandEncoder,
339 camera: &CameraParams,
340 ) {
341 let mut pc = pass.with(&pipeline.draw_pipeline);
342 pc.bind(
343 0,
344 &DrawData {
345 draw_particles: self.particle_buf.into(),
346 camera: *camera,
347 },
348 );
349 pc.draw(0, 4, 0, self.capacity as u32);
350 }
351}
352
353fn pack_color(c: [u8; 4]) -> u32 {
354 (c[0] as u32) | ((c[1] as u32) << 8) | ((c[2] as u32) << 16) | ((c[3] as u32) << 24)
355}