diff --git a/examples/testbed/src/main.rs b/examples/testbed/src/main.rs index a84ac19..eaa931a 100644 --- a/examples/testbed/src/main.rs +++ b/examples/testbed/src/main.rs @@ -127,7 +127,7 @@ async fn main() { light_tran.rotate_y(math::Angle::Degrees(25.0)); world.spawn(( DirectionalLight { - enabled: true, + enabled: false, color: Vec3::ONE, intensity: 0.35 //..Default::default() @@ -174,6 +174,25 @@ async fn main() { ), cube_mesh.clone(), )); + + /* world.spawn(( + SpotLight { + enabled: true, + color: Vec3::new(1.0, 0.0, 0.0), + intensity: 1.0, + range: 1.5, + ..Default::default() + }, + Transform::new( + Vec3::new(0.0, 0.2, -1.5), + //Vec3::new(-5.0, 1.0, -0.28), + //Vec3::new(-10.0, 0.94, -0.28), + + Quat::IDENTITY, + Vec3::new(0.15, 0.15, 0.15), + ), + cube_mesh.clone(), + )); */ } /* { diff --git a/lyra-game/src/render/light/mod.rs b/lyra-game/src/render/light/mod.rs index 7d94e2e..ca37007 100644 --- a/lyra-game/src/render/light/mod.rs +++ b/lyra-game/src/render/light/mod.rs @@ -6,7 +6,7 @@ use lyra_ecs::{Entity, Tick, World, query::{Entities, TickOf}}; pub use point::*; pub use spotlight::*; -use std::{collections::{HashMap, VecDeque}, marker::PhantomData}; +use std::{collections::{HashMap, VecDeque}, marker::PhantomData, mem}; use crate::math::Transform; @@ -103,13 +103,14 @@ pub(crate) struct LightUniformBuffers { pub buffer: wgpu::Buffer, pub bind_group_pair: BindGroupPair, pub light_indexes: HashMap, + dead_indices: VecDeque, pub current_light_idx: u32, } impl LightUniformBuffers { pub fn new(device: &wgpu::Device) -> Self { let limits = device.limits(); - // TODO: check this limit somehow + // TODO: ensure we dont write over this limit let max_buffer_sizes = (limits.max_uniform_buffer_binding_size as u64) / 2; let buffer = device.create_buffer( @@ -161,45 +162,88 @@ impl LightUniformBuffers { bind_group_pair: BindGroupPair::new(bindgroup, bindgroup_layout), light_indexes: Default::default(), current_light_idx: 0, + dead_indices: VecDeque::new(), } } + /// Returns the index for the entity, and if this index is new + fn get_index_for(&mut self, missed: &mut HashMap, entity: Entity) -> (bool, u32) { + let idx = missed.remove(&entity) + .map(|v| (false, v)) + .or_else(|| + self.dead_indices.pop_front() + .map(|v| (true, v)) + ) + .unwrap_or_else(|| { + let t = self.current_light_idx; + self.current_light_idx += 1; + (true, t) + }); + idx + } + pub fn update_lights(&mut self, queue: &wgpu::Queue, world_tick: Tick, world: &World) { - let mut lights = LightsUniform::default(); + // used to detect what lights were removed + let mut missed_lights: HashMap = self.light_indexes.drain().collect(); for (entity, point_light, transform, light_epoch, transform_epoch) in world.view_iter::<(Entities, &PointLight, &Transform, TickOf, TickOf)>() { - // TODO: dont update light every frame - let idx = *self.light_indexes.entry(entity) - .or_insert_with(|| { - let t = self.current_light_idx; - self.current_light_idx += 1; - t - }) as usize; + let (new, idx) = self.get_index_for(&mut missed_lights, entity); + self.light_indexes.insert(entity, idx); - let uniform = LightUniform::from_point_light_bundle(&point_light, &transform); - lights.data[idx] = uniform; + if new || light_epoch == world_tick || transform_epoch == world_tick { + let uniform = LightUniform::from_point_light_bundle(&point_light, &transform); + + let offset = mem::size_of::() * 4 + mem::size_of::() * idx as usize; + queue.write_buffer(&self.buffer, offset as _, bytemuck::cast_slice(&[uniform])); + } } - if let Some((entity, dir_light, transform)) = - world.view_iter::<(Entities, &DirectionalLight, &Transform)>().next() { + for (entity, spot_light, transform, light_epoch, transform_epoch) + in world.view_iter::<(Entities, &SpotLight, &Transform, TickOf, TickOf)>() { + + let (new, idx) = self.get_index_for(&mut missed_lights, entity); + self.light_indexes.insert(entity, idx); + + if new || light_epoch == world_tick || transform_epoch == world_tick { + let uniform = LightUniform::from_spot_light_bundle(&spot_light, &transform); + + let offset = mem::size_of::() * 4 + mem::size_of::() * idx as usize; + queue.write_buffer(&self.buffer, offset as _, bytemuck::cast_slice(&[uniform])); + } + } + + for (entity, dir_light, transform, light_epoch, transform_epoch) + in world.view_iter::<(Entities, &DirectionalLight, &Transform, TickOf, TickOf)>() { - let idx = *self.light_indexes.entry(entity) - .or_insert_with(|| { - let t = self.current_light_idx; - self.current_light_idx += 1; - t - }) as usize; - - let uniform = LightUniform::from_directional_bundle(&dir_light, &transform); - lights.data[idx] = uniform; + let (new, idx) = self.get_index_for(&mut missed_lights, entity); + self.light_indexes.insert(entity, idx); + + if new || light_epoch == world_tick || transform_epoch == world_tick { + let uniform = LightUniform::from_directional_bundle(&dir_light, &transform); + let offset = mem::size_of::() * 4 + mem::size_of::() * idx as usize; + queue.write_buffer(&self.buffer, offset as _, bytemuck::cast_slice(&[uniform])); + } } - lights.light_count = self.light_indexes.len() as u32; + // anything left in missed_lights were lights that were deleted + let len = missed_lights.len(); + self.dead_indices.reserve(len); - // update the light count in the struct - queue.write_buffer(&self.buffer, 0, bytemuck::cast_slice(&[lights])); + for (_, v) in missed_lights.drain() { + // write zeros in place of this now dead light, the enabled boolean will be set to false + let mut zeros = Vec::new(); + zeros.resize(mem::size_of::(), 0u32); + + let offset = mem::size_of::() * 4 + mem::size_of::() * v as usize; + queue.write_buffer(&self.buffer, offset as _, bytemuck::cast_slice(zeros.as_slice())); + + self.dead_indices.push_back(v); + } + + // update the amount of lights, then the array of lights + queue.write_buffer(&self.buffer, 0, bytemuck::cast_slice(&[self.current_light_idx as u32])); } } @@ -227,8 +271,8 @@ pub(crate) struct LightUniform { pub intensity: f32, pub smoothness: f32, - pub spot_cutoff: f32, - pub spot_outer_cutoff: f32, + pub spot_cutoff_rad: f32, + pub spot_outer_cutoff_rad: f32, } impl LightUniform { @@ -244,8 +288,8 @@ impl LightUniform { intensity: light.intensity, smoothness: light.smoothness, - spot_cutoff: 0.0, - spot_outer_cutoff: 0.0, + spot_cutoff_rad: 0.0, + spot_outer_cutoff_rad: 0.0, } } @@ -262,190 +306,27 @@ impl LightUniform { intensity: light.intensity, smoothness: 0.0, - spot_cutoff: 0.0, - spot_outer_cutoff: 0.0, + spot_cutoff_rad: 0.0, + spot_outer_cutoff_rad: 0.0, } } // Create the SpotLightUniform from an ECS bundle - /* pub fn from_bundle(light: &SpotLight, transform: &Transform) -> Self { + pub fn from_spot_light_bundle(light: &SpotLight, transform: &Transform) -> Self { Self { + light_type: LightType::Spotlight as u32, + enabled: light.enabled as u32, position: transform.translation, - _padding: 0, direction: transform.forward(), - _padding2: 0, color: light.color, - cutoff: light.cutoff.to_radians().cos(), - outer_cutoff: light.outer_cutoff.to_radians().cos(), - constant: light.constant, - linear: light.linear, - quadratic: light.quadratic, - ambient: light.ambient, - diffuse: light.diffuse, - specular: light.specular, - _padding3: 0, - } - } */ -} -#[repr(C)] -#[derive(Default, Debug, Copy, Clone, bytemuck::Pod, bytemuck::Zeroable)] -pub struct LightsUniform { - light_count: u32, - _padding: [u32; 3], - data: [LightUniform; 10], // TODO: make this a dynamic length -} - -/* #[repr(C)] -#[derive(Default, Debug, Copy, Clone, bytemuck::Pod, bytemuck::Zeroable)] -pub struct LightsUniform { - light_count: u32, - _padding: [u32; 3], - spot_lights: [SpotLightUniform; MAX_LIGHT_COUNT], - spot_light_count: u32, - _padding2: [u32; 3], - directional_light: DirectionalLightUniform, -} - -#[repr(C)] -#[derive(Default, Debug, Copy, Clone, bytemuck::Pod, bytemuck::Zeroable)] -pub struct PointLightUniform { - /// The position of the light - /// vec4 is used here for gpu padding, w is ignored in the shader - pub position: glam::Vec4, - - /// The color of the light - /// vec4 is used here for gpu padding, w is ignored in the shader - pub color: glam::Vec4, - - /// The intensity of the light - /// This works by just multiplying the result of the lighting - /// calculations by this scalar - pub intensity: f32, - - /// The constant used in the quadratic attenuation calculation. Its best to leave this at 1.0 - pub constant: f32, - - /// The linear factor used in the quadratic attenuation calculation. - pub linear: f32, - - /// The quadratic factor used in the quadratic attenuation calculation. - pub quadratic: f32, - - pub ambient: f32, - pub diffuse: f32, - pub specular: f32, - - pub _padding: u32, -} - -impl PointLightUniform { - /// Create the PointLightUniform from an ECS bundle - pub fn from_bundle(light: &PointLight, transform: &Transform) -> Self { - Self { - position: glam::Vec4::new(transform.translation.x, transform.translation.y, transform.translation.z, 0.0), - //_padding: 0, - color: glam::Vec4::new(light.color.x, light.color.y, light.color.z, 0.0), - //_padding2: 0, + range: light.range, intensity: light.intensity, - constant: light.constant, - linear: light.linear, - quadratic: light.quadratic, + smoothness: light.smoothness, - ambient: light.ambient, - diffuse: light.diffuse, - specular: light.specular, - - _padding: 0, + spot_cutoff_rad: light.cutoff.to_radians(), + spot_outer_cutoff_rad: light.outer_cutoff.to_radians(), } } } -#[repr(C)] -#[derive(Default, Debug, Copy, Clone, bytemuck::Pod, bytemuck::Zeroable)] -pub struct DirectionalLightUniform { - /// The direction of the light - pub direction: glam::Vec3, - // gpu padding - pub _padding: u32, - /// The color of the light - pub color: glam::Vec3, - // no padding is needed here since ambient acts as the padding - // that would usually be needed for the vec3 - - /// The scalar of the ambient light created by this caster. - pub ambient: f32, - /// The scalar of the diffuse light created by this caster. - pub diffuse: f32, - /// The scalar of the specular reflections created by this caster. - pub specular: f32, - - pub _padding2: [u32; 2], -} - -impl DirectionalLightUniform { - /// Create the DirectionalLightUniform from an ECS bundle - pub fn from_bundle(light: &DirectionalLight, transform: &Transform) -> Self { - //transform.forward() - Self { - direction: transform.forward(), - _padding: 0, - color: light.color, - ambient: light.ambient, - diffuse: light.diffuse, - specular: light.specular, - _padding2: [0; 2], - } - } -} - -#[repr(C)] -#[derive(Default, Debug, Copy, Clone, bytemuck::Pod, bytemuck::Zeroable)] -pub(crate) struct SpotLightUniform { - pub position: glam::Vec3, - pub _padding: u32, - pub direction: glam::Vec3, - pub _padding2: u32, - pub color: glam::Vec3, - // no padding is needed here since cutoff acts as the padding - // that would usually be needed for the vec3 - - pub cutoff: f32, - pub outer_cutoff: f32, - - /// The constant used in the quadratic attenuation calculation. Its best to leave this at 1.0 - pub constant: f32, - - /// The linear factor used in the quadratic attenuation calculation. - pub linear: f32, - - /// The quadratic factor used in the quadratic attenuation calculation. - pub quadratic: f32, - - pub ambient: f32, - pub diffuse: f32, - pub specular: f32, - pub _padding3: u32, -} - -impl SpotLightUniform { - /// Create the SpotLightUniform from an ECS bundle - pub fn from_bundle(light: &SpotLight, transform: &Transform) -> Self { - Self { - position: transform.translation, - _padding: 0, - direction: transform.forward(), - _padding2: 0, - color: light.color, - cutoff: light.cutoff.to_radians().cos(), - outer_cutoff: light.outer_cutoff.to_radians().cos(), - constant: light.constant, - linear: light.linear, - quadratic: light.quadratic, - ambient: light.ambient, - diffuse: light.diffuse, - specular: light.specular, - _padding3: 0, - } - } -} */ \ No newline at end of file diff --git a/lyra-game/src/render/light/spotlight.rs b/lyra-game/src/render/light/spotlight.rs index c0eeed8..fa89c5a 100644 --- a/lyra-game/src/render/light/spotlight.rs +++ b/lyra-game/src/render/light/spotlight.rs @@ -4,15 +4,26 @@ use crate::math; #[derive(Debug, Clone, Component)] pub struct SpotLight { + pub enabled: bool, pub color: glam::Vec3, + pub range: f32, + pub intensity: f32, + pub smoothness: f32, pub cutoff: math::Angle, pub outer_cutoff: math::Angle, +} - pub constant: f32, - pub linear: f32, - pub quadratic: f32, +impl Default for SpotLight { + fn default() -> Self { + Self { + enabled: true, + color: glam::Vec3::new(1.0, 1.0, 1.0), + range: 1.0, + intensity: 1.0, + smoothness: 0.75, - pub ambient: f32, - pub diffuse: f32, - pub specular: f32, + cutoff: math::Angle::Degrees(45.0), + outer_cutoff: math::Angle::Degrees(45.0), + } + } } \ No newline at end of file diff --git a/lyra-game/src/render/light_cull_compute.rs b/lyra-game/src/render/light_cull_compute.rs index 98b742e..66fd876 100644 --- a/lyra-game/src/render/light_cull_compute.rs +++ b/lyra-game/src/render/light_cull_compute.rs @@ -83,7 +83,6 @@ impl LightCullCompute { label: Some("BGL_LightIndicesGrid"), }); - // TODO: resize texture when screen is resized let size = wgpu::Extent3d { width: workgroup_size.x, height: workgroup_size.y, @@ -168,7 +167,8 @@ impl LightCullCompute { source: wgpu::ShaderSource::Wgsl(Cow::Borrowed(shader_src)), }); - let workgroup_size = glam::UVec2::new((screen_size.width as f32 / 16.0).ceil() as u32, (screen_size.height as f32 / 16.0).ceil() as u32); + let workgroup_size = glam::UVec2::new((screen_size.width as f32 / 16.0).ceil() as u32, + (screen_size.height as f32 / 16.0).ceil() as u32); let light_grid = Self::create_grid(&device, workgroup_size); let depth_tex_pair = depth_texture.create_bind_group(&device); @@ -205,7 +205,11 @@ impl LightCullCompute { pub fn update_screen_size(&mut self, size: PhysicalSize) { self.screen_size_buffer.write_buffer(&self.queue, 0, &[UVec2::new(size.width, size.height)]); - self.workgroup_size = glam::UVec2::new((size.width as f32 / 16.0).ceil() as u32, (size.height as f32 / 16.0).ceil() as u32); + self.workgroup_size = glam::UVec2::new((size.width as f32 / 16.0).ceil() as u32, + (size.height as f32 / 16.0).ceil() as u32); + + // I hate that the entire bind group is recreated on a resize but its the only way :( + self.light_indices_grid = Self::create_grid(&self.device, self.workgroup_size); } pub fn compute(&mut self, camera_buffers: &BufferWrapper, lights_buffers: &LightUniformBuffers, depth_texture: &RenderTexture) { @@ -228,6 +232,7 @@ impl LightCullCompute { pass.dispatch_workgroups(self.workgroup_size.x, self.workgroup_size.y, 1); } + self.queue.submit(std::iter::once(encoder.finish())); self.device.poll(wgpu::Maintain::Wait); self.cleanup(); diff --git a/lyra-game/src/render/shaders/base.wgsl b/lyra-game/src/render/shaders/base.wgsl index aa907a7..7b97ddf 100755 --- a/lyra-game/src/render/shaders/base.wgsl +++ b/lyra-game/src/render/shaders/base.wgsl @@ -238,5 +238,50 @@ fn blinn_phong_point_light(world_pos: vec3, world_norm: vec3, point_li } fn blinn_phong_spot_light(world_pos: vec3, world_norm: vec3, spot_light: Light, material: Material, specular_factor: vec3) -> vec3 { - return vec3(0.0); // TODO + let light_color = spot_light.color; + let light_pos = spot_light.position; + let camera_view_pos = u_camera.position; + + let light_dir = normalize(spot_light.position - world_pos); + + var ambient_color = light_color * material.ambient.xyz * material.diffuse.xyz; + + //// diffuse //// + //let light_dir = normalize(light_pos - world_pos); + + let diffuse_strength = max(dot(world_norm, light_dir), 0.0); + var diffuse_color = light_color * (diffuse_strength * material.diffuse.xyz); + //// end of diffuse //// + + //// specular //// + let view_dir = normalize(camera_view_pos - world_pos); + let half_dir = normalize(view_dir + light_dir); + + let specular_strength = pow(max(dot(world_norm, half_dir), 0.0), material.shininess); + var specular_color = specular_strength * (light_color * specular_factor); + //// end of specular //// + + //// spot light soft edges //// + let theta = dot(light_dir, normalize(-spot_light.direction)); + let epsilon = spot_light.spot_cutoff - spot_light.spot_outer_cutoff; + let intensity = clamp((theta - spot_light.spot_outer_cutoff) / epsilon, 0.0, 1.0); + //diffuse_color *= intensity; + //specular_color *= intensity; + //// end of spot light soft edges //// + + //// spot light attenuation //// + let distance = length(light_pos - world_pos); + let attenuation = calc_attenuation(spot_light, distance); + + ambient_color *= attenuation * intensity; + diffuse_color *= attenuation * intensity; + specular_color *= attenuation * intensity; + //// end of spot light attenuation //// + + + return /*ambient_color +*/ diffuse_color + specular_color; +} + +fn calc_attenuation(light: Light, distance: f32) -> f32 { + return 1.0 - smoothstep(light.range * light.smoothness, light.range, distance); } \ No newline at end of file diff --git a/lyra-game/src/render/shaders/light_cull.comp.wgsl b/lyra-game/src/render/shaders/light_cull.comp.wgsl index d8e481b..dfe5a7a 100644 --- a/lyra-game/src/render/shaders/light_cull.comp.wgsl +++ b/lyra-game/src/render/shaders/light_cull.comp.wgsl @@ -33,6 +33,13 @@ struct Lights { data: array, }; +struct Cone { + tip: vec3, + height: f32, + direction: vec3, + radius: f32, +} + var wg_min_depth: atomic; var wg_max_depth: atomic; var wg_light_index_start: atomic; @@ -160,13 +167,22 @@ fn cs_main( && sphere_inside_frustrum(wg_frustum_planes, position, radius)) { // TODO: add the light to the transparent geometry list - add_light(light_index); - if (!sphere_inside_plane(position, radius, wg_frustum_planes[4])) { - + add_light(light_index); + } + } else if (light.light_ty == LIGHT_TY_SPOT) { + let dir_vs = u_camera.view * vec4(light.direction, 1.0); + let cone_radius = tan(light.spot_cutoff) * light.range; + let cone = Cone(position, radius, dir_vs.xyz, cone_radius); + + if (cone_inside_frustum(cone, wg_frustum_planes)) { + // TODO: add the light to the transparent geometry list + + if (!cone_inside_plane(cone, wg_frustum_planes[4])) { + add_light(light_index); + } } } - // TODO: spotlights } } @@ -258,4 +274,45 @@ fn compute_plane(p0: vec3, p1: vec3, p2: vec3) -> vec4 { plane.w = dot(plane.xyz, p0); return plane; +} + +fn point_inside_plane(point: vec3, plane: vec4) -> bool { + return dot(plane.xyz, point) - plane.w < 0.0; +} + +/// Check to see if a cone if fully behind (inside the negative halfspace of) a plane. +/// +/// Source: Real-time collision detection, Christer Ericson (2005) +/// (https://www.3dgep.com/forward-plus/#light-culling-compute-shader) +fn cone_inside_plane(cone: Cone, plane: vec4) -> bool { + // Compute the farthest point on the end of the cone to the positive space of the plane. + let m = cross(cross(plane.xyz, cone.direction), cone.direction); + let farthest = cone.tip + cone.direction * cone.height - m * cone.radius; + + // The cone is in the negative halfspace of the plane if the tip of the cone, + // and the farthest point on the end of the cone are inside the negative halfspace + // of the plane. + return point_inside_plane(cone.tip, plane) && point_inside_plane(farthest, plane); +} + +fn cone_inside_frustum(cone: Cone, frustum: array, 6>) -> bool { + //let near_plane = frustum[4]; + //let far_plane = frustum[5]; + + // check near and far clipping planes first + //if (cone_inside_plane(cone, near_plane) || cone_inside_plane(cone, far_plane)) { + // return false; + //} + + // to be able to index this array with a non-const value, + // it must be defined as a var + var frustum_v = frustum; + + for (var i = 0u; i < 4u; i++) { + if (cone_inside_plane(cone, frustum_v[i])) { + return false; + } + } + + return true; } \ No newline at end of file diff --git a/shell.nix b/shell.nix index d0c65e9..98db997 100755 --- a/shell.nix +++ b/shell.nix @@ -15,13 +15,19 @@ in mold udev lua5_4_compat - (nixpkgs.rustChannelOf { rustToolchain = ./rust-toolchain.toml; }).rust + ((nixpkgs.rustChannelOf { rustToolchain = ./rust-toolchain.toml; }).rust.override { + extensions = [ + "rust-src" + "rust-analysis" + ]; + }) ]; buildInputs = [ udev alsa-lib libGL gcc vulkan-loader vulkan-headers vulkan-tools xorg.libX11 xorg.libXcursor xorg.libXi xorg.libXrandr # To use the x11 feature libxkbcommon wayland # To use the wayland feature + ]; LD_LIBRARY_PATH = lib.makeLibraryPath buildInputs; }