From 445f1c454ca7fabbebdbf40102316f843c1b0fb0 Mon Sep 17 00:00:00 2001 From: Albin Chaboissier Date: Tue, 8 Sep 2026 15:07:36 +0200 Subject: [PATCH] =?UTF-8?q?To=20mouni=C3=A9?= MIME-Version: 1.0 Content-Type: text/plain; charset=UTF-8 Content-Transfer-Encoding: 8bit --- Cargo.toml | 1 + shaders/voxel.wgsl | 139 +++++++----- src/host_production.rs | 15 ++ src/main.rs | 349 +++++++++++++++++++++++++---- src/producers.rs | 4 +- src/voxel_cache/structure_table.rs | 5 +- 6 files changed, 418 insertions(+), 95 deletions(-) create mode 100644 src/host_production.rs diff --git a/Cargo.toml b/Cargo.toml index cf25611..00b2760 100644 --- a/Cargo.toml +++ b/Cargo.toml @@ -16,6 +16,7 @@ glam = "0.33.5" image = "0.25.10" indicatif = "0.18.6" itertools = "0.15.0" +ordered-float = "5.5.0" pollster = "1.0.1" rand = "0.10.2" rayon = "1.12.0" diff --git a/shaders/voxel.wgsl b/shaders/voxel.wgsl index 5381edc..1bfb5fd 100644 --- a/shaders/voxel.wgsl +++ b/shaders/voxel.wgsl @@ -73,11 +73,6 @@ fn unpack_color(color: u32) -> vec4 @group(0) @binding(5) var structure_table_pointer: array; @group(0) @binding(6) var structure_table_request_buffer: array>; -struct FragmentOutput { - @location(0) color: vec4, - @builtin(frag_depth) depth: f32, // Equivalent to gl_FragDepth -} - @vertex fn chunk(@builtin(vertex_index) index: u32, @location(0) position: vec3, @location(1) id: u32) -> VertexOutput { @@ -221,10 +216,10 @@ fn new_traverse(ray_dir: vec3, ray_origin: vec3, root_id: u32, dist_of let cone_factor = tan(fovy_rad / 2.) * 2.; let st_pointer = structure_table_pointer[root_id]; - if (!node_subdivided(st_pointer)) { + discard; var result: HitResult; result.color = vec4(0., 1., 0., 1.); result.hit_pos = ray_origin; @@ -236,6 +231,7 @@ fn new_traverse(ray_dir: vec3, ray_origin: vec3, root_id: u32, dist_of // Send request on structure table atomicAdd(&structure_table_request_buffer[root_id], 1); + discard; var result: HitResult; result.color = vec4(0., 1., 0., 1.); result.hit_pos = ray_origin; @@ -270,7 +266,9 @@ fn new_traverse(ray_dir: vec3, ray_origin: vec3, root_id: u32, dist_of child_size = 1 << u32(node_shift - 2); var child_pos = (voxel >> vec3(u32(node_shift - 2))) & vec3(3); - var pointer = structure_pool[dfs_stack[current_depth]].pointers[child_pos.x + child_pos.y * 4 + child_pos.z * 4 * 4]; + var child_index = child_pos.x + child_pos.y * 4 + child_pos.z * 4 * 4; + var current_node = dfs_stack[current_depth]; + var pointer = structure_pool[current_node].pointers[child_index]; let min_child_size = (length(vec3(voxel) - pos_origin) + dist_offset_voxel) * cone_factor; while(node_subdivided(pointer) && @@ -280,7 +278,7 @@ fn new_traverse(ray_dir: vec3, ray_origin: vec3, root_id: u32, dist_of if(!node_pointer_valid(pointer) && node_subdivided(pointer)) { // Record request - atomicAdd(&request_buffer[dfs_stack[current_depth]].requests[child_pos.x + child_pos.y * 4 + child_pos.z * 4 * 4], 1); + atomicAdd(&request_buffer[current_node].requests[child_pos.x + child_pos.y * 4 + child_pos.z * 4 * 4], 1); break; } @@ -290,21 +288,24 @@ fn new_traverse(ray_dir: vec3, ray_origin: vec3, root_id: u32, dist_of node_shift = (max_depth - current_depth) * 2; child_size = 1 << u32(node_shift - 2); child_pos = (voxel >> vec3(u32(node_shift - 2))) & vec3(3); - dfs_stack[current_depth] = node_pointer(pointer); + child_index = child_pos.x + child_pos.y * 4 + child_pos.z * 4 * 4; + current_node = node_pointer(pointer); + dfs_stack[current_depth] = current_node; - pointer = structure_pool[dfs_stack[current_depth]].pointers[child_pos.x + child_pos.y * 4 + child_pos.z * 4 * 4]; + pointer = structure_pool[current_node].pointers[child_index]; // Record usage - usage_buffer[dfs_stack[current_depth]] = constants.frame_timestamp; + usage_buffer[current_node] = constants.frame_timestamp; } // Check color - let color = color_pool[dfs_stack[current_depth]].colors[child_pos.x + child_pos.y * 4 + child_pos.z * 4 * 4]; + let color = color_pool[current_node].colors[child_index]; if(((color >> 24) & 0xFF) != 0) { var result: HitResult; result.color = unpack_color(color); + //result.color = vec4(f32(iter) / 400.); result.hit_pos = (far_t / f32(1 << u32(max_depth * 2))) * ray_dir + ray_origin; return result; } @@ -350,17 +351,90 @@ fn new_traverse(ray_dir: vec3, ray_origin: vec3, root_id: u32, dist_of return result; } +struct FragmentOutput { + @location(0) color: vec4, + @builtin(frag_depth) depth: f32, // Equivalent to gl_FragDepth +} + +struct FragmentPrepassOutput { + @builtin(frag_depth) depth: f32, // Equivalent to gl_FragDepth +} + +@early_depth_test(less_equal) +@fragment +fn fragment_prepass(in: VertexOutput) -> FragmentPrepassOutput +{ + let ray_dir = normalize(in.world_pos - in.cam_pos); + let interp = box_inter(in.cam_pos - in.chunk_position, ray_dir, vec3(0.), vec3(1)); + let ray_origin = (in.cam_pos - in.chunk_position) + ray_dir * (max(0., interp.x)); + + let result = new_traverse(ray_dir, ray_origin, in.structure_id, length(in.cam_pos - (ray_origin + in.chunk_position))); + let clip_pos = constants.view_proj * vec4(result.hit_pos + in.chunk_position, 1.); + let depth = clip_pos.z / clip_pos.w; + var frag_out: FragmentPrepassOutput; + + //frag_out.color = result.color; + //frag_out.color = result.color; + + frag_out.depth = depth; + return frag_out; +} + +@early_depth_test(less_equal) +@fragment +fn fragment(in: VertexOutput) -> FragmentOutput +{ + //frag_out.color = vec4(2 * 0.01 / (100. + 0.01 - depth * (100. - 0.01))); + let ray_dir = normalize(in.world_pos - in.cam_pos); + let interp = box_inter(in.cam_pos - in.chunk_position, ray_dir, vec3(0.), vec3(1)); + let ray_origin = (in.cam_pos - in.chunk_position) + ray_dir * (max(0., interp.x)); + + + let result = new_traverse(ray_dir, ray_origin, in.structure_id, length(in.cam_pos - (ray_origin + in.chunk_position))); + let clip_pos = constants.view_proj * vec4(result.hit_pos + in.chunk_position, 1.); + let depth = clip_pos.z / clip_pos.w; + var frag_out: FragmentOutput; + //frag_out.color = result.color; + frag_out.color = result.color; + frag_out.depth = depth; + return frag_out; + + //return vec4(ray_origin, 1.); + //return frag_out; + //return vec4(interp.y / 10.); +} + +/* +@fragment +fn fragment(in: VertexOutput) -> @location(0) vec4 +{ + let st = structure_table_pointer[0]; + let subdivided = ((st >> 31) & 1) != 0; + let pointer_valid = ((st >> 30) & 1) != 0; + // Request stuff + atomicAdd(&structure_table_request_buffer[0], 1); + if(subdivided && !pointer_valid) + { + return vec4(0., 1., 0., 1.); + } + return vec4(1., 0., 0., 1.); +} +*/ + +/* fn traverse(ray_dir: vec3, ray_origin: vec3, root_id: u32, dist_offset: f32) -> vec4 { let st_pointer = structure_table_pointer[root_id]; if (!node_subdivided(st_pointer)) { + discard; return vec4(0., 1., 0., 1.); } if(!node_pointer_valid(st_pointer)) { atomicAdd(&structure_table_request_buffer[root_id], 1); + discard; return vec4(0., 1., 0., 1.); } @@ -547,45 +621,4 @@ fn traverse(ray_dir: vec3, ray_origin: vec3, root_id: u32, dist_offset return vec4(1., 0., 1., 1.); } - -@early_depth_test(less_equal) -@fragment -fn fragment(in: VertexOutput) -> FragmentOutput -{ - //frag_out.color = vec4(2 * 0.01 / (100. + 0.01 - depth * (100. - 0.01))); - let ray_dir = normalize(in.world_pos - in.cam_pos); - let interp = box_inter(in.cam_pos - in.chunk_position, ray_dir, vec3(0.), vec3(1)); - let ray_origin = (in.cam_pos - in.chunk_position) + ray_dir * (max(0., interp.x)); - - - let result = new_traverse(ray_dir, ray_origin, in.structure_id, length(in.cam_pos - (ray_origin + in.chunk_position))); - let clip_pos = constants.view_proj * vec4(result.hit_pos + in.chunk_position, 1.); - let depth = clip_pos.z / clip_pos.w; - var frag_out: FragmentOutput; - //frag_out.color = result.color; - frag_out.color = result.color; - frag_out.depth = depth; - return frag_out; - - //return vec4(ray_origin, 1.); - //return frag_out; - //return vec4(interp.y / 10.); -} - -/* -@fragment -fn fragment(in: VertexOutput) -> @location(0) vec4 -{ - let st = structure_table_pointer[0]; - let subdivided = ((st >> 31) & 1) != 0; - let pointer_valid = ((st >> 30) & 1) != 0; - // Request stuff - atomicAdd(&structure_table_request_buffer[0], 1); - if(subdivided && !pointer_valid) - { - return vec4(0., 1., 0., 1.); - } - return vec4(1., 0., 0., 1.); -} */ - diff --git a/src/host_production.rs b/src/host_production.rs new file mode 100644 index 0000000..d385f25 --- /dev/null +++ b/src/host_production.rs @@ -0,0 +1,15 @@ +// Contains useful facilities to render data streamed in from the host + +// Can produce a voxel given +// - Its depth +// - Its position within the chunk +// - The chunks position +pub trait ChunkVoxelProducer +{ + fn produce_voxel( + &mut self, + depth: usize, + chunk_position: (usize, usize, usize), + voxel_position: (usize, usize, usize), + ); +} diff --git a/src/main.rs b/src/main.rs index 1eb9efe..0db87be 100644 --- a/src/main.rs +++ b/src/main.rs @@ -16,11 +16,15 @@ use glam::Mat4; use glam::Vec3; use glam::Vec4; use itertools::Itertools; +use ordered_float::OrderedFloat; use rayon::iter::IndexedParallelIterator; use rayon::iter::IntoParallelRefIterator; use rayon::iter::ParallelIterator; +use wgpu::BindGroupLayout; +use wgpu::BlendState; use wgpu::Buffer; use wgpu::BufferUsages; +use wgpu::ComputePipeline; use wgpu::Device; use wgpu::Extent3d; use wgpu::Features; @@ -28,7 +32,9 @@ use wgpu::InstanceDescriptor; use wgpu::InstanceFlags; use wgpu::MemoryBudgetThresholds; use wgpu::Operations; +use wgpu::PipelineCompilationOptions; use wgpu::RenderPipeline; +use wgpu::ShaderStages; use wgpu::Texture; use wgpu::TextureUsages; use wgpu::TextureView; @@ -44,6 +50,7 @@ use winit::event_loop::ActiveEventLoop; use winit::event_loop::ControlFlow; use winit::event_loop::EventLoop; use winit::event_loop::OwnedDisplayHandle; +use winit::platform::x11::EventLoopBuilderExtX11; use winit::window::Window; use winit::window::WindowId; @@ -64,6 +71,7 @@ use crate::voxel_cache::producer_interface::CacheRequest; mod camera; mod egui_renderer; +mod host_production; mod producers; mod sparse_tree; mod voxel_cache; @@ -77,11 +85,16 @@ struct State queue: wgpu::Queue, size: winit::dpi::PhysicalSize, surface: wgpu::Surface<'static>, - depth_buffer: (wgpu::Texture, wgpu::TextureView), + depth_buffer: (wgpu::Texture, wgpu::TextureView, wgpu::TextureView), + prepass_depth_buffer: (wgpu::Texture, wgpu::TextureView, wgpu::TextureView), + prepass_downsampling: u32, + upsample_pipeline: ComputePipeline, surface_format: wgpu::TextureFormat, egui_renderer: EguiRenderer, pipeline: RenderPipeline, + prepass_pipeline: RenderPipeline, + prepass_upsample_bg_layout: BindGroupLayout, voxel_cache: Arc>>, cache_interface: Arc>, terrain_generator: Arc>, @@ -158,10 +171,12 @@ impl State let egui_renderer = EguiRenderer::new(&device, surface_format, &window); - let mut voxel_cache = VoxelCache::<4>::new(200_000, device.clone(), queue.clone()); + let mut voxel_cache = VoxelCache::<4>::new(100_000, device.clone(), queue.clone()); let cache_interface = CacheProducerInterface::new(1024, &device); - let terrain_generator = TerrainGenerator::<4>::new(5, "vxls_height.tif", 0.2, "img.jpg"); + let terrain_generator = + TerrainGenerator::<4>::new(5, "vxls_height.tif", 0.2, "img_low.jpg"); + //let terrain_generator = TerrainGenerator::<4>::new(5, "vxls_height.tif", 0.2, "img.jpg"); // let terrain_generator = TerrainGenerator::<4>::new( // 5, // "./pointe_percee/height.tif", @@ -170,9 +185,9 @@ impl State // ); // let terrain_generator = TerrainGenerator::<4>::new( // 5, - // "/home/albin/Documents/vxls_maps/lapiz/height.tif", + // "/home/albin/Documents/vxls_maps/orgere/height.tif", // 0.2, - // "/home/albin/Documents/vxls_maps/lapiz/ortho.jpg", + // "/home/albin/Documents/vxls_maps/orgere/color.jpg", // ); let mut chunk_pos_map = HashMap::new(); @@ -216,6 +231,61 @@ impl State immediate_size: size_of::() as u32, }); + let prepass_pipeline = device.create_render_pipeline(&wgpu::RenderPipelineDescriptor { + label: Some("Render pipeline"), + layout: Some(&pipeline_layout), + vertex: wgpu::VertexState { + module: &shader_module, + entry_point: Some("chunk"), + compilation_options: Default::default(), + buffers: &[Some(wgpu::VertexBufferLayout { + array_stride: (size_of::() * 3 + size_of::()) as u64, + step_mode: wgpu::VertexStepMode::Instance, + attributes: &[ + wgpu::VertexAttribute { + format: wgpu::VertexFormat::Float32x3, + offset: 0, + shader_location: 0, + }, + wgpu::VertexAttribute { + format: wgpu::VertexFormat::Uint32, + offset: (size_of::() * 3) as u64, + shader_location: 1, + }, + ], + })], + }, + primitive: wgpu::PrimitiveState { + topology: wgpu::PrimitiveTopology::TriangleList, + strip_index_format: None, + front_face: wgpu::FrontFace::Ccw, + cull_mode: Some(wgpu::Face::Front), + unclipped_depth: false, + polygon_mode: wgpu::PolygonMode::Fill, + conservative: false, + }, + depth_stencil: Some(wgpu::DepthStencilState { + format: wgpu::TextureFormat::Depth24PlusStencil8, + depth_write_enabled: Some(true), + depth_compare: Some(wgpu::CompareFunction::Less), + stencil: wgpu::StencilState::default(), + bias: wgpu::DepthBiasState::default(), + }), + multisample: wgpu::MultisampleState::default(), + fragment: Some(wgpu::FragmentState { + module: &shader_module, + entry_point: Some("fragment_prepass"), + compilation_options: wgpu::PipelineCompilationOptions::default(), + targets: &[Some(wgpu::ColorTargetState { + format: surface_format, + blend: None, + write_mask: wgpu::ColorWrites::default(), + })], + }), + multiview_mask: None, + cache: None, + }); + let chunk_pipeline = device.create_render_pipeline(&wgpu::RenderPipelineDescriptor { label: Some("Render pipeline"), layout: Some(&pipeline_layout), @@ -271,6 +341,82 @@ impl State cache: None, }); + let prepass_upsample_bg_layout = + device.create_bind_group_layout(&wgpu::BindGroupLayoutDescriptor { + label: Some("prepass_upsample_bg"), + entries: &[ + wgpu::BindGroupLayoutEntry { + binding: 0, + visibility: ShaderStages::COMPUTE, + ty: wgpu::BindingType::StorageTexture { + access: wgpu::StorageTextureAccess::ReadOnly, + format: wgpu::TextureFormat::Depth32Float, + view_dimension: wgpu::TextureViewDimension::D2, + }, + count: None, + }, + wgpu::BindGroupLayoutEntry { + binding: 1, + visibility: ShaderStages::COMPUTE, + ty: wgpu::BindingType::StorageTexture { + access: wgpu::StorageTextureAccess::WriteOnly, + format: wgpu::TextureFormat::Depth24PlusStencil8, + view_dimension: wgpu::TextureViewDimension::D2, + }, + count: None, + }, + ], + }); + + let prepass_upsample_layout = + device.create_pipeline_layout(&wgpu::PipelineLayoutDescriptor { + label: Some("prepass_upsample_layout"), + bind_group_layouts: &[Some(&prepass_upsample_bg_layout)], + immediate_size: 0, + }); + + let prepass_downsampling = 4; + let prepass_upsample = device.create_compute_pipeline(&wgpu::ComputePipelineDescriptor { + label: Some("prepass_upsample"), + layout: Some(&prepass_upsample_layout), + module: &device.create_shader_module(wgpu::ShaderModuleDescriptor { + label: Some("prepass_upsample_module"), + source: wgpu::ShaderSource::Wgsl( + format!( + " + @group(0) @binding(0) var input_tex: texture_storage_2d; + @group(0) @binding(1) var output_tex: texture_storage_2d; + @compute @workgroup_size(8, 8) + fn main(@builtin(global_invocation_id) id: vec3) + {{ + let dp = vec2(id.xy); + let sp = dp * {prepass_downsampling}; + + var depth = 1.; + for(var ox = 0; ox < {prepass_downsampling}; ox ++) + {{ + for(var oy = 0; oy < {prepass_downsampling}; oy ++) + {{ + depth = min(depth, textureLoad(sp + vec(ox, oy), p, 0)); + }} + }} + + textureStore( + output_tex, + dp, + depth + ); + }} + " + ) + .into(), + ), + }), + entry_point: Some("main"), + compilation_options: PipelineCompilationOptions::default(), + cache: None, + }); + let state = State { instance, window, @@ -279,13 +425,22 @@ impl State surface_format, egui_renderer, depth_buffer: Self::create_depth_buffer(&device, size.width, size.height), + prepass_depth_buffer: Self::create_depth_buffer( + &device, + size.width / prepass_downsampling, + size.height / prepass_downsampling, + ), + prepass_downsampling, + upsample_pipeline: prepass_upsample, queue, device, usage_vec: Arc::new(Mutex::new(vec![])), pipeline: chunk_pipeline, + prepass_pipeline, + prepass_upsample_bg_layout, voxel_cache: Arc::new(Mutex::new(voxel_cache)), cache_interface: cache_interface.into(), - insertion_debounce: false, + insertion_debounce: true, camera: Default::default(), instance_buffer, instance_count, @@ -307,6 +462,19 @@ impl State fn handle_event(&mut self, event: &WindowEvent) { + if let WindowEvent::KeyboardInput { event, .. } = event + { + match (event.state, event.physical_key) + { + ( + winit::event::ElementState::Pressed, + winit::keyboard::PhysicalKey::Code(winit::keyboard::KeyCode::KeyF), + ) => self.insertion_debounce = !self.insertion_debounce, + _ => + {} + } + } + self.egui_renderer.handle_input(&self.window, event); self.camera.handle_input(event); } @@ -326,7 +494,11 @@ impl State } } - fn create_depth_buffer(device: &Device, width: u32, height: u32) -> (Texture, TextureView) + fn create_depth_buffer( + device: &Device, + width: u32, + height: u32, + ) -> (Texture, TextureView, TextureView) { let texture = device.create_texture(&wgpu::wgt::TextureDescriptor { label: Some("Depth buffer"), @@ -346,7 +518,12 @@ impl State label: Some("depth view"), ..Default::default() }); - (texture, texture_view) + let texture_view_upsample = texture.create_view(&wgpu::wgt::TextureViewDescriptor { + label: Some("prepass depth view upsample"), + aspect: wgpu::TextureAspect::DepthOnly, + ..Default::default() + }); + (texture, texture_view, texture_view_upsample) } fn configure_surface(&self) @@ -375,6 +552,11 @@ impl State self.configure_surface(); self.depth_buffer = Self::create_depth_buffer(&self.device, new_size.width, new_size.height); + self.prepass_depth_buffer = Self::create_depth_buffer( + &self.device, + new_size.width / self.prepass_downsampling, + new_size.height / self.prepass_downsampling, + ); } fn render(&mut self) @@ -382,6 +564,34 @@ impl State self.camera.update(); self.voxel_cache.lock().next_frame(); + // Build sorted buffer + let mut chunks = self.chunk_pos_map.iter().collect::>(); + + chunks.sort_by_key(|(_, (x, y, z))| { + OrderedFloat( + (Vec3::new(*x as f32, *y as f32, *z as f32) - self.camera.position).length(), + ) + }); + + let instance_buffer = self + .device + .create_buffer_init(&wgpu::util::BufferInitDescriptor { + label: Some("Instance buffer"), + contents: bytemuck::cast_slice( + chunks + .iter() + .map(|(structure_id, (x, y, z))| InstanceAttribute { + x: *x as f32, + y: *y as f32, + z: *z as f32, + id: **structure_id, + }) + .collect::>() + .as_slice(), + ), + usage: BufferUsages::COPY_DST | BufferUsages::VERTEX, + }); + // Create texture view. // NOTE: We must handle Timeout because the surface may be unavailable // (e.g., when the window is occluded on macOS). @@ -436,8 +646,73 @@ impl State mapped_at_creation: false, }); - // ~~ Main render pass ~~ + // ~~ Prepass upsample bind group ~~ + let prepass_bg = self.device.create_bind_group(&wgpu::BindGroupDescriptor { + label: Some("prepass_bind_group"), + layout: &self.prepass_upsample_bg_layout, + entries: &[ + wgpu::BindGroupEntry { + binding: 0, + resource: wgpu::BindingResource::TextureView(&self.prepass_depth_buffer.2), + }, + wgpu::BindGroupEntry { + binding: 1, + resource: wgpu::BindingResource::TextureView(&self.depth_buffer.2), + }, + ], + }); + let mut encoder = self.device.create_command_encoder(&Default::default()); + // ~~ Prepass ~~ + { + let mut renderpass = encoder.begin_render_pass(&wgpu::RenderPassDescriptor { + label: None, + color_attachments: &[], + depth_stencil_attachment: Some(wgpu::RenderPassDepthStencilAttachment { + view: &self.prepass_depth_buffer.1, + depth_ops: Some(Operations { + load: wgpu::LoadOp::Clear(1.), + store: wgpu::StoreOp::Store, + }), + stencil_ops: None, + }), + timestamp_writes: None, + occlusion_query_set: None, + multiview_mask: None, + }); + renderpass.set_vertex_buffer(0, instance_buffer.slice(..)); + renderpass.set_bind_group(0, Some(&self.voxel_cache.lock().bind_group()), &[]); + let imm = [Immediates { + view_proj: self.camera.view_proj(), + cam_pos: self.camera.position, + frame_timestamp: self.voxel_cache.lock().current_timestamp(), + }]; + renderpass.set_immediates(0, unsafe { as_raw_bytes(&imm) }); + + renderpass.set_pipeline(&self.prepass_pipeline); + renderpass.draw(0..36, 0..(self.instance_count as u32)); + + // End the renderpass. + drop(renderpass); + } + + // ~~ Upsample pass ~~ + { + let mut compute_pass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor { + label: Some("upsample_compute_pass"), + timestamp_writes: None, + }); + + compute_pass.set_bind_group(0, Some(&prepass_bg), &[]); + compute_pass.set_pipeline(&self.upsample_pipeline); + compute_pass.dispatch_workgroups( + (self.size.width / self.prepass_downsampling).div_ceil(8), + (self.size.height / self.prepass_downsampling).div_ceil(8), + 1, + ); + } + + // ~~ Main render pass ~~ { let mut renderpass = encoder.begin_render_pass(&wgpu::RenderPassDescriptor { label: None, @@ -471,9 +746,7 @@ impl State occlusion_query_set: None, multiview_mask: None, }); - - renderpass.set_pipeline(&self.pipeline); - renderpass.set_vertex_buffer(0, self.instance_buffer.slice(..)); + renderpass.set_vertex_buffer(0, instance_buffer.slice(..)); renderpass.set_bind_group(0, Some(&self.voxel_cache.lock().bind_group()), &[]); let imm = [Immediates { view_proj: self.camera.view_proj(), @@ -481,6 +754,8 @@ impl State frame_timestamp: self.voxel_cache.lock().current_timestamp(), }]; renderpass.set_immediates(0, unsafe { as_raw_bytes(&imm) }); + + renderpass.set_pipeline(&self.pipeline); renderpass.draw(0..36, 0..(self.instance_count as u32)); // End the renderpass. @@ -492,13 +767,10 @@ impl State // ~~ EGUI Render pass ~~ { self.egui_renderer.begin_frame(&self.window); - egui::Window::new("Window ! ").resizable(true).show( + egui::Window::new("Window ! ").resizable(false).show( self.egui_renderer.context(), |ui| { - if self - .camera - .pressed_keyset - .contains(&winit::keyboard::KeyCode::KeyF) + if !self.insertion_debounce { ui.label( egui::RichText::new("Cache paused") @@ -561,6 +833,23 @@ impl State }, ); + // ~~ Do cache managment + // if !self + // .camera + // .pressed_keyset + // .contains(&winit::keyboard::KeyCode::KeyF) + if self.insertion_debounce + { + self.voxel_cache + .lock() + .cache_post_render(&mut encoder, &self.cache_interface); + } + + // ~~ Submit command buffer ~~ + self.queue.submit([encoder.finish()]); + self.window.pre_present_notify(); + self.queue.present(surface_texture); + // ~~ Get Ray-marching timestamps, report time ~~ let cloned_rm_time = self.rm_time.clone(); let cloned_queue = self.queue.clone(); @@ -577,28 +866,12 @@ impl State ); // ~~ Do cache managment - if !self - .camera - .pressed_keyset - .contains(&winit::keyboard::KeyCode::KeyF) + // if !self + // .camera + // .pressed_keyset + // .contains(&winit::keyboard::KeyCode::KeyF) + if self.insertion_debounce { - self.voxel_cache - .lock() - .cache_post_render(&mut encoder, &self.cache_interface); - } - - // ~~ Submit command buffer ~~ - self.queue.submit([encoder.finish()]); - self.window.pre_present_notify(); - self.queue.present(surface_texture); - - // ~~ Do cache managment - if !self - .camera - .pressed_keyset - .contains(&winit::keyboard::KeyCode::KeyF) - { - self.insertion_debounce = true; let request_count = self .cache_interface .total_request_count(&self.device, &self.queue); diff --git a/src/producers.rs b/src/producers.rs index 96079b4..6ad10c1 100644 --- a/src/producers.rs +++ b/src/producers.rs @@ -3,6 +3,8 @@ use std::path::Path; use glam::Vec3; use itertools::Itertools; +use rayon::iter::IntoParallelRefMutIterator; +use rayon::iter::ParallelIterator; use crate::sparse_tree::Color; use crate::voxel_cache::data::ExplicitNTreeNode; @@ -310,7 +312,7 @@ where let heightmap_max = heightmap.iter().copied().reduce(f32::max).unwrap(); heightmap - .iter_mut() + .par_iter_mut() .filter(|x| **x == -9999.) .for_each(|x| *x = heightmap_min); diff --git a/src/voxel_cache/structure_table.rs b/src/voxel_cache/structure_table.rs index c66179b..d643c69 100644 --- a/src/voxel_cache/structure_table.rs +++ b/src/voxel_cache/structure_table.rs @@ -80,7 +80,7 @@ impl StructureTable &mut self.request_buffer } - pub fn remove_structure(&mut self, structure_id: u32) + pub fn free_structure(&mut self, structure_id: u32) { self.available_slots += 1; self.allocation_table[structure_id as usize] = false; @@ -104,8 +104,7 @@ impl StructureTable .allocation_table .iter() .enumerate() - .filter(|(_, allocated)| !**allocated) - .next() + .find(|(_, allocated)| !**allocated) .unwrap(); self.allocation_table[first_id] = true; self.available_slots -= 1;