diff --git a/shared/src/core/bssrdf.rs b/shared/src/core/bssrdf.rs index 25c249b..fa755cc 100644 --- a/shared/src/core/bssrdf.rs +++ b/shared/src/core/bssrdf.rs @@ -6,7 +6,7 @@ use crate::spectra::{N_SPECTRUM_SAMPLES, SampledSpectrum}; use crate::utils::Ptr; use crate::utils::math::{catmull_rom_weights, square}; use crate::utils::sampling::sample_catmull_rom_2d; -use crate::utils::{DevicePtr, ptr::Slice}; +use crate::utils::{Ptr, ptr::Slice}; use crate::{Float, PI}; use enum_dispatch::enum_dispatch; use std::sync::Arc; @@ -78,15 +78,15 @@ impl From<&SubsurfaceInteraction> for SurfaceInteraction { dndv: Normal3f::zero(), }, face_index: 0, - area_light: DevicePtr::null(), - material: DevicePtr::null(), + area_light: Ptr::null(), + material: Ptr::null(), dpdx: Vector3f::zero(), dpdy: Vector3f::zero(), dudx: 0., dvdx: 0., dudy: 0., dvdy: 0., - shape: DevicePtr::from(&Shape::default()), + shape: Ptr::from(&Shape::default()), } } } @@ -96,11 +96,11 @@ impl From<&SubsurfaceInteraction> for SurfaceInteraction { pub struct BSSRDFTable { pub n_rho_samples: u32, pub n_radius_samples: u32, - pub rho_samples: DevicePtr, - pub radius_samples: DevicePtr, - pub profile: DevicePtr, - pub rho_eff: DevicePtr, - pub profile_cdf: DevicePtr, + pub rho_samples: Ptr, + pub radius_samples: Ptr, + pub profile: Ptr, + pub rho_eff: Ptr, + pub profile_cdf: Ptr, } impl BSSRDFTable { @@ -165,7 +165,7 @@ pub struct TabulatedBSSRDF { eta: Float, sigma_t: SampledSpectrum, rho: SampledSpectrum, - table: DevicePtr, + table: Ptr, } impl TabulatedBSSRDF { @@ -187,7 +187,7 @@ impl TabulatedBSSRDF { eta, sigma_t, rho, - table: DevicePtr::from(table), + table: Ptr::from(table), } } diff --git a/shared/src/core/color.rs b/shared/src/core/color.rs index bb78711..6e81558 100644 --- a/shared/src/core/color.rs +++ b/shared/src/core/color.rs @@ -4,14 +4,16 @@ use std::ops::{ Add, AddAssign, Div, DivAssign, Index, IndexMut, Mul, MulAssign, Neg, Sub, SubAssign, }; +use crate::Float; use crate::core::geometry::Point2f; -use crate::core::pbrt::{Float, find_interval}; use crate::core::spectrum::Spectrum; +use crate::utils::find_interval; use crate::utils::math::{SquareMatrix, SquareMatrix3f, clamp, evaluate_polynomial, lerp}; use enum_dispatch::enum_dispatch; -#[derive(Debug, Clone)] +#[repr(C)] +#[derive(Debug, Default, Clone, Copy)] pub struct XYZ { pub x: Float, pub y: Float, @@ -24,6 +26,12 @@ impl From<(Float, Float, Float)> for XYZ { } } +impl From<[Float; 3]> for XYZ { + fn from(triplet: (Float, Float, Float)) -> Self { + XYZ::new(triplet.0, triplet.1, triplet.2) + } +} + impl<'a> IntoIterator for &'a XYZ { type Item = &'a Float; type IntoIter = std::array::IntoIter<&'a Float, 3>; @@ -81,9 +89,9 @@ impl XYZ { } } -impl Index for XYZ { +impl Index for XYZ { type Output = Float; - fn index(&self, index: usize) -> &Self::Output { + fn index(&self, index: u32) -> &Self::Output { debug_assert!(index < 3); match index { 0 => &self.x, @@ -93,8 +101,8 @@ impl Index for XYZ { } } -impl IndexMut for XYZ { - fn index_mut(&mut self, index: usize) -> &mut Self::Output { +impl IndexMut for XYZ { + fn index_mut(&mut self, index: u32) -> &mut Self::Output { debug_assert!(index < 3); match index { 0 => &mut self.x, @@ -247,7 +255,8 @@ impl fmt::Display for XYZ { } } -#[derive(Debug, Default, Copy, Clone)] +#[repr(C)] +#[derive(Debug, Default, Clone, Copy)] pub struct RGB { pub r: Float, pub g: Float, @@ -286,7 +295,7 @@ impl RGB { self.r.min(self.g).min(self.b) } - pub fn min_component_index(&self) -> usize { + pub fn min_component_index(&self) -> u32 { if self.r < self.g { if self.r < self.b { 0 } else { 2 } } else { @@ -294,7 +303,7 @@ impl RGB { } } - pub fn max_component_index(&self) -> usize { + pub fn max_component_index(&self) -> u32 { if self.r > self.g { if self.r > self.b { 0 } else { 2 } } else { @@ -315,9 +324,9 @@ impl RGB { } } -impl Index for RGB { +impl Index for RGB { type Output = Float; - fn index(&self, index: usize) -> &Self::Output { + fn index(&self, index: u32) -> &Self::Output { debug_assert!(index < 3); match index { 0 => &self.r, @@ -327,8 +336,8 @@ impl Index for RGB { } } -impl IndexMut for RGB { - fn index_mut(&mut self, index: usize) -> &mut Self::Output { +impl IndexMut for RGB { + fn index_mut(&mut self, index: u32) -> &mut Self::Output { debug_assert!(index < 3); match index { 0 => &mut self.r, @@ -658,7 +667,7 @@ impl ColorEncodingTrait for SRGBEncoding { fn to_linear(&self, vin: &[u8], vout: &mut [Float]) { for (i, &v) in vin.iter().enumerate() { - vout[i] = SRGB_TO_LINEAR_LUT[v as usize]; + vout[i] = SRGB_TO_LINEAR_LUT[v as u32]; } } @@ -961,7 +970,7 @@ const SRGB_TO_LINEAR_LUT: [Float; 256] = [ 1.0000000000, ]; -pub const RES: usize = 64; +pub const RES: u32 = 64; #[repr(C)] #[derive(Clone, Copy, Debug, Default)] @@ -1007,9 +1016,9 @@ unsafe impl Sync for RGBToSpectrumTable {} impl RGBToSpectrumTable { #[inline(always)] - fn get_coeffs(&self, bucket: usize, z: usize, y: usize, x: usize) -> Coeffs { + fn get_coeffs(&self, bucket: u32, z: u32, y: u32, x: u32) -> Coeffs { let offset = bucket * (RES * RES * RES) + z * (RES * RES) + y * (RES) + x; - unsafe { *self.coeffs.add(offset) } + unsafe { *self.coeffs.add(offset as usize) } } pub fn evaluate(&self, rgb: RGB) -> RGBSigmoidPolynomial { @@ -1045,25 +1054,25 @@ impl RGBToSpectrumTable { let x = coord_a / z; let y = coord_b / z; - let z_nodes_slice = unsafe { core::slice::from_raw_parts(self.z_nodes, RES) }; - let zi = find_interval(RES, |i| z_nodes_slice[i] < z); + let z_nodes_slice = unsafe { core::slice::from_raw_parts(self.z_nodes, RES as usize) }; + let zi = find_interval(RES, |i| z_nodes_slice[i as usize] < z) as usize; let dz = (z - z_nodes_slice[zi]) / (z_nodes_slice[zi + 1] - z_nodes_slice[zi]); let x_float = x * (RES - 1) as Float; - let xi = (x_float as usize).min(RES - 2); + let xi = (x_float as u32).min(RES - 2); let dx = x_float - xi as Float; let y_float = y * (RES - 1) as Float; - let yi = (y_float as usize).min(RES - 2); + let yi = (y_float as u32).min(RES - 2); let dy = y_float - yi as Float; - let c000 = self.get_coeffs(c_idx, zi, yi, xi); - let c001 = self.get_coeffs(c_idx, zi, yi, xi + 1); - let c010 = self.get_coeffs(c_idx, zi, yi + 1, xi); - let c011 = self.get_coeffs(c_idx, zi, yi + 1, xi + 1); - let c100 = self.get_coeffs(c_idx, zi + 1, yi, xi); - let c101 = self.get_coeffs(c_idx, zi + 1, yi, xi + 1); - let c110 = self.get_coeffs(c_idx, zi + 1, yi + 1, xi); - let c111 = self.get_coeffs(c_idx, zi + 1, yi + 1, xi + 1); + let c000 = self.get_coeffs(c_idx, zi as u32, yi, xi); + let c001 = self.get_coeffs(c_idx, zi as u32, yi, xi + 1); + let c010 = self.get_coeffs(c_idx, zi as u32, yi + 1, xi); + let c011 = self.get_coeffs(c_idx, zi as u32, yi + 1, xi + 1); + let c100 = self.get_coeffs(c_idx, zi as u32 + 1, yi, xi); + let c101 = self.get_coeffs(c_idx, zi as u32 + 1, yi, xi + 1); + let c110 = self.get_coeffs(c_idx, zi as u32 + 1, yi + 1, xi); + let c111 = self.get_coeffs(c_idx, zi as u32 + 1, yi + 1, xi + 1); let c00 = lerp(dx, c000, c001); let c01 = lerp(dx, c010, c011); let c10 = lerp(dx, c100, c101); diff --git a/shared/src/core/interaction.rs b/shared/src/core/interaction.rs index eb438c9..0d5cc25 100644 --- a/shared/src/core/interaction.rs +++ b/shared/src/core/interaction.rs @@ -16,7 +16,7 @@ use crate::core::sampler::{Sampler, SamplerTrait}; use crate::core::shape::Shape; use crate::core::texture::{GPUFloatTexture, UniversalTextureEvaluator}; use crate::spectra::{SampledSpectrum, SampledWavelengths}; -use crate::utils::DevicePtr; +use crate::utils::Ptr; use crate::utils::math::{clamp, difference_of_products, square}; use enum_dispatch::enum_dispatch; use std::any::Any; @@ -31,7 +31,7 @@ pub struct InteractionBase { pub wo: Vector3f, pub uv: Point2f, pub medium_interface: MediumInterface, - pub medium: DevicePtr, + pub medium: Ptr, } impl InteractionBase { @@ -49,11 +49,11 @@ impl InteractionBase { wo: wo.normalize(), time, medium_interface: MediumInterface::default(), - medium: DevicePtr::null(), + medium: Ptr::null(), } } - pub fn new_medium(p: Point3f, wo: Vector3f, time: Float, medium: DevicePtr) -> Self { + pub fn new_medium(p: Point3f, wo: Vector3f, time: Float, medium: Ptr) -> Self { Self { pi: Point3fi::new_from_point(p), n: Normal3f::zero(), @@ -114,7 +114,7 @@ pub trait InteractionTrait { false } - fn get_medium(&self, w: Vector3f) -> DevicePtr { + fn get_medium(&self, w: Vector3f) -> Ptr { let data = self.get_common(); if !data.medium_interface.inside.is_null() || !data.medium_interface.outside.is_null() { if w.dot(data.n.into()) > 0.0 { @@ -225,9 +225,9 @@ pub struct SurfaceInteraction { pub dndv: Normal3f, pub shading: ShadingGeom, pub face_index: u32, - pub area_light: DevicePtr, - pub material: DevicePtr, - pub shape: DevicePtr, + pub area_light: Ptr, + pub material: Ptr, + pub shape: Ptr, pub dpdx: Vector3f, pub dpdy: Vector3f, pub dudx: Float, @@ -412,8 +412,8 @@ impl SurfaceInteraction { fn compute_bump_geom( &mut self, tex_eval: &UniversalTextureEvaluator, - displacement: DevicePtr, - normal_image: DevicePtr, + displacement: Ptr, + normal_image: Ptr, ) { let ctx = NormalBumpEvalContext::from(&*self); let (dpdu, dpdv) = if !displacement.is_null() { @@ -536,7 +536,7 @@ impl InteractionTrait for SurfaceInteraction { &mut self.common } - fn get_medium(&self, w: Vector3f) -> DevicePtr { + fn get_medium(&self, w: Vector3f) -> Ptr { let interface = self.common.medium_interface; if self.n().dot(w.into()) > 0.0 { interface.outside @@ -583,16 +583,16 @@ impl SurfaceInteraction { dndu, dndv, }, - material: DevicePtr::null(), + material: Ptr::null(), face_index: 0, - area_light: DevicePtr::null(), + area_light: Ptr::null(), dpdx: Vector3f::zero(), dpdy: Vector3f::zero(), dudx: 0.0, dudy: 0.0, dvdx: 0.0, dvdy: 0.0, - shape: DevicePtr::null(), + shape: Ptr::null(), } } @@ -658,13 +658,13 @@ impl SurfaceInteraction { ray_medium: &Medium, prim_medium_interface: MediumInterface, ) { - self.material = DevicePtr::from(mtl); - self.area_light = DevicePtr::from(area); + self.material = Ptr::from(mtl); + self.area_light = Ptr::from(area); if prim_medium_interface.is_medium_transition() { self.common.medium_interface = prim_medium_interface; } else { - self.common.medium = DevicePtr::from(ray_medium); + self.common.medium = Ptr::from(ray_medium); } } } @@ -681,7 +681,7 @@ impl MediumInteraction { p: Point3f, wo: Vector3f, time: Float, - medium: DevicePtr, + medium: Ptr, phase: PhaseFunction, ) -> Self { Self { diff --git a/shared/src/core/light.rs b/shared/src/core/light.rs index 32089b0..c5bab66 100644 --- a/shared/src/core/light.rs +++ b/shared/src/core/light.rs @@ -340,9 +340,9 @@ pub enum Light { DiffuseArea(DiffuseAreaLight), Distant(DistantLight), Goniometric(GoniometricLight), - InfiniteUniform(InfiniteUniformLight), - InfiniteImage(InfiniteImageLight), - InfinitePortal(InfinitePortalLight), + InfiniteUniform(UniformInfiniteLight), + InfiniteImage(ImageInfiniteLight), + InfinitePortal(PortalInfiniteLight), Point(PointLight), Projection(ProjectionLight), Spot(SpotLight), diff --git a/shared/src/core/pbrt.rs b/shared/src/core/pbrt.rs index ae43b1d..99ceeff 100644 --- a/shared/src/core/pbrt.rs +++ b/shared/src/core/pbrt.rs @@ -5,8 +5,50 @@ use std::hash::Hash; use std::ops::{Add, Mul}; use std::sync::{Arc, Mutex}; +use crate::core::image::DeviceImage; +use crate::core::light::LightTrait; +use crate::core::shape::Shape; +use crate::core::texture::GPUFloatTexture; +use crate::lights::*; +use crate::spectra::{DenselySampledSpectrum, RGBColorSpace}; +use crate::utils::Ptr; + pub type Float = f32; +// #[derive(Copy, Clone, Debug)] +// pub struct Host; +// +// #[derive(Copy, Clone, Debug)] +// pub struct Device; +// +// pub trait Backend: Copy + Clone + 'static { +// type ShapeRef: Copy + Clone; +// type TextureRef: Copy + Clone; +// type ImageRef: Copy + Clone; +// type DenseSpectrumRef = Ptr; +// type ColorSpaceRef = Ptr; +// type DiffuseLight: LightTrait; +// type PointLight: LightTrait; +// type UniformInfiniteLight: LightTrait; +// type PortalInfiniteLight: LightTrait; +// type ImageInfiniteLight: LightTrait; +// type SpotLight: LightTrait; +// } +// +// impl Backend for Device { +// type ShapeRef = Ptr; +// type TextureRef = Ptr; +// type ColorSpaceRef = Ptr; +// type DenseSpectrumRef = Ptr; +// type ImageRef = Ptr; +// type DiffuseLight = Ptr>; +// type PointLight = Ptr>; +// type UniformInfiniteLight = Ptr>; +// type PortalInfiniteLight = Ptr>; +// type ImageInfiniteLight = Ptr>; +// type SpotLight = Ptr>; +// } +// #[cfg(not(feature = "use_f64"))] pub type FloatBits = u32; @@ -100,36 +142,11 @@ pub const PI_OVER_2: Float = 1.570_796_326_794_896_619_23; pub const PI_OVER_4: Float = 0.785_398_163_397_448_309_61; pub const SQRT_2: Float = 1.414_213_562_373_095_048_80; -#[inline] -pub fn find_interval(sz: u32, pred: F) -> u32 -where - F: Fn(u32) -> bool, -{ - let mut first = 0; - let mut len = sz; - - while len > 0 { - let half = len >> 1; - let middle = first + half; - - if pred(middle) { - first = middle + 1; - len -= half + 1; - } else { - len = half; - } - } - - let ret = (first as i32 - 1).max(0) as u32; - ret.min(sz.saturating_sub(2)) -} - #[inline] pub fn gamma(n: i32) -> Float { n as Float * MACHINE_EPSILON / (1. - n as Float * MACHINE_EPSILON) } -// Define the static counters. These are thread-safe. pub static RARE_EVENT_TOTAL_CALLS: AtomicU64 = AtomicU64::new(0); pub static RARE_EVENT_CONDITION_MET: AtomicU64 = AtomicU64::new(0); diff --git a/shared/src/core/sampler.rs b/shared/src/core/sampler.rs index 441fdb5..89cc6e7 100644 --- a/shared/src/core/sampler.rs +++ b/shared/src/core/sampler.rs @@ -1,8 +1,8 @@ use crate::core::filter::FilterTrait; use crate::core::geometry::{Bounds2f, Point2f, Point2i, Vector2f}; use crate::core::options::{PBRTOptions, get_options}; -use crate::core::pbrt::{Float, ONE_MINUS_EPSILON, PI, PI_OVER_2, PI_OVER_4, find_interval}; -use crate::utils::DevicePtr; +use crate::core::pbrt::{Float, ONE_MINUS_EPSILON, PI, PI_OVER_2, PI_OVER_4}; +use crate::utils::Ptr; use crate::utils::containers::Array2D; use crate::utils::math::{ BinaryPermuteScrambler, DigitPermutation, FastOwenScrambler, NoRandomizer, OwenScrambler, @@ -103,7 +103,7 @@ pub struct HaltonSampler { mult_inverse: [u64; 2], halton_index: u64, dim: u32, - digit_permutations: DevicePtr, + digit_permutations: Ptr, } impl HaltonSampler { diff --git a/shared/src/lib.rs b/shared/src/lib.rs index 89aa05b..d37d345 100644 --- a/shared/src/lib.rs +++ b/shared/src/lib.rs @@ -1,6 +1,7 @@ #![allow(unused_imports, dead_code)] #![feature(float_erf)] #![feature(f16)] +#![feature(associated_type_defaults)] pub mod bxdfs; pub mod cameras; diff --git a/shared/src/lights/diffuse.rs b/shared/src/lights/diffuse.rs index 8dfcc1f..270ec8c 100644 --- a/shared/src/lights/diffuse.rs +++ b/shared/src/lights/diffuse.rs @@ -1,4 +1,3 @@ -use crate::PI; use crate::core::color::{RGB, XYZ}; use crate::core::geometry::*; use crate::core::image::{DeviceImage, ImageAccess}; @@ -9,7 +8,6 @@ use crate::core::light::{ LightBase, LightBounds, LightLiSample, LightSampleContext, LightTrait, LightType, }; use crate::core::medium::MediumInterface; -use crate::core::pbrt::Float; use crate::core::shape::{Shape, ShapeSampleContext, ShapeTrait}; use crate::core::spectrum::{Spectrum, SpectrumTrait}; use crate::core::texture::{ @@ -17,17 +15,18 @@ use crate::core::texture::{ }; use crate::spectra::*; use crate::utils::hash::hash_float; -use crate::utils::{DevicePtr, Transform}; +use crate::utils::{Ptr, Transform}; +use crate::{Float, PI}; #[repr(C)] #[derive(Clone, Debug, Copy)] pub struct DiffuseAreaLight { pub base: LightBase, - pub shape: DevicePtr, - pub alpha: DevicePtr, - pub image_color_space: DevicePtr, - pub lemit: DevicePtr, - pub image: DevicePtr, + pub shape: Ptr, + pub alpha: Ptr, + pub colorspace: Ptr, + pub lemit: Ptr, + pub image: Ptr, pub area: Float, pub two_sided: bool, pub scale: Float, diff --git a/shared/src/lights/distant.rs b/shared/src/lights/distant.rs index 2d1def4..070097a 100644 --- a/shared/src/lights/distant.rs +++ b/shared/src/lights/distant.rs @@ -5,14 +5,14 @@ use crate::core::interaction::{Interaction, InteractionBase, SimpleInteraction}; use crate::core::light::{LightBase, LightBounds, LightLiSample, LightSampleContext, LightTrait}; use crate::core::spectrum::SpectrumTrait; use crate::spectra::{DenselySampledSpectrum, SampledSpectrum, SampledWavelengths}; -use crate::utils::{DevicePtr, Ptr}; +use crate::utils::Ptr; use crate::{Float, PI}; #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct DistantLight { pub base: LightBase, - pub lemit: DevicePtr, + pub lemit: Ptr, pub scale: Float, pub scene_center: Point3f, pub scene_radius: Float, diff --git a/shared/src/lights/goniometric.rs b/shared/src/lights/goniometric.rs index 323cc6d..75054cf 100644 --- a/shared/src/lights/goniometric.rs +++ b/shared/src/lights/goniometric.rs @@ -8,7 +8,7 @@ use crate::core::spectrum::{Spectrum, SpectrumTrait}; use crate::spectra::{DenselySampledSpectrum, SampledSpectrum, SampledWavelengths}; use crate::utils::math::equal_area_sphere_to_square; use crate::utils::sampling::DevicePiecewiseConstant2D; -use crate::utils::{DevicePtr, Transform}; +use crate::utils::{Ptr, Transform}; use crate::{Float, PI}; #[derive(Debug, Clone, Copy)] @@ -16,8 +16,8 @@ pub struct GoniometricLight { pub base: LightBase, pub iemit: DenselySampledSpectrum, pub scale: Float, - pub image: DevicePtr, - pub distrib: DevicePtr, + pub image: Ptr, + pub distrib: Ptr, } impl GoniometricLight { diff --git a/shared/src/lights/infinite.rs b/shared/src/lights/infinite.rs index 4a9ea76..8595cf7 100644 --- a/shared/src/lights/infinite.rs +++ b/shared/src/lights/infinite.rs @@ -13,30 +13,29 @@ use crate::core::medium::{Medium, MediumInterface}; use crate::core::spectrum::{Spectrum, SpectrumTrait}; use crate::spectra::{DenselySampledSpectrum, SampledSpectrum, SampledWavelengths}; use crate::spectra::{RGBColorSpace, RGBIlluminantSpectrum}; -use crate::utils::Transform; use crate::utils::math::{clamp, equal_area_sphere_to_square, equal_area_square_to_sphere, square}; -use crate::utils::ptr::DevicePtr; use crate::utils::sampling::{ AliasTable, DevicePiecewiseConstant2D, WindowedPiecewiseConstant2D, sample_uniform_sphere, uniform_sphere_pdf, }; +use crate::utils::{Ptr, Transform}; use crate::{Float, PI}; use std::sync::Arc; #[repr(C)] #[derive(Debug, Copy, Clone)] -pub struct InfiniteUniformLight { +pub struct UniformInfiniteLight { pub base: LightBase, - pub lemit: DevicePtr, + pub lemit: Ptr, pub scale: Float, pub scene_center: Point3f, pub scene_radius: Float, } -unsafe impl Send for InfiniteUniformLight {} -unsafe impl Sync for InfiniteUniformLight {} +unsafe impl Send for UniformInfiniteLight {} +unsafe impl Sync for UniformInfiniteLight {} -impl LightTrait for InfiniteUniformLight { +impl LightTrait for UniformInfiniteLight { fn base(&self) -> &LightBase { &self.base } @@ -113,21 +112,21 @@ impl LightTrait for InfiniteUniformLight { #[repr(C)] #[derive(Clone, Copy, Debug)] -pub struct InfiniteImageLight { +pub struct ImageInfiniteLight { pub base: LightBase, - pub image: DevicePtr, - pub image_color_space: DevicePtr, - pub distrib: DevicePtr, - pub compensated_distrib: DevicePtr, + pub image: Ptr, + pub image_color_space: Ptr, + pub distrib: Ptr, + pub compensated_distrib: Ptr, pub scale: Float, pub scene_radius: Float, pub scene_center: Point3f, } -unsafe impl Send for InfiniteImageLight {} -unsafe impl Sync for InfiniteImageLight {} +unsafe impl Send for ImageInfiniteLight {} +unsafe impl Sync for ImageInfiniteLight {} -impl InfiniteImageLight { +impl ImageInfiniteLight { fn image_le(&self, uv: Point2f, lambda: &SampledWavelengths) -> SampledSpectrum { let mut rgb = RGB::default(); for c in 0..3 { @@ -142,7 +141,7 @@ impl InfiniteImageLight { } } -impl LightTrait for InfiniteImageLight { +impl LightTrait for ImageInfiniteLight { fn base(&self) -> &LightBase { &self.base } @@ -249,10 +248,10 @@ impl LightTrait for InfiniteImageLight { #[repr(C)] #[derive(Debug, Copy, Clone)] -pub struct InfinitePortalLight { +pub struct PortalInfiniteLight { pub base: LightBase, - pub image: DevicePtr, - pub image_color_space: DevicePtr, + pub image: Ptr, + pub image_color_space: Ptr, pub scale: Float, pub portal: [Point3f; 4], pub portal_frame: Frame, @@ -261,7 +260,7 @@ pub struct InfinitePortalLight { pub scene_radius: Float, } -impl InfinitePortalLight { +impl PortalInfiniteLight { pub fn image_lookup(&self, uv: Point2f, lambda: &SampledWavelengths) -> SampledSpectrum { let mut rgb = RGB::default(); for c in 0..3 { @@ -318,7 +317,7 @@ impl InfinitePortalLight { } } -impl LightTrait for InfinitePortalLight { +impl LightTrait for PortalInfiniteLight { fn base(&self) -> &LightBase { &self.base } diff --git a/shared/src/lights/mod.rs b/shared/src/lights/mod.rs index 4006aaf..f670f85 100644 --- a/shared/src/lights/mod.rs +++ b/shared/src/lights/mod.rs @@ -10,7 +10,7 @@ pub mod spot; pub use diffuse::DiffuseAreaLight; pub use distant::DistantLight; pub use goniometric::GoniometricLight; -pub use infinite::{InfiniteImageLight, InfinitePortalLight, InfiniteUniformLight}; +pub use infinite::{ImageInfiniteLight, PortalInfiniteLight, UniformInfiniteLight}; pub use point::PointLight; pub use projection::ProjectionLight; pub use spot::SpotLight; diff --git a/shared/src/lights/point.rs b/shared/src/lights/point.rs index 18c7f19..c2daa2a 100644 --- a/shared/src/lights/point.rs +++ b/shared/src/lights/point.rs @@ -7,7 +7,7 @@ use crate::core::light::{ }; use crate::core::spectrum::SpectrumTrait; use crate::spectra::{DenselySampledSpectrum, SampledSpectrum, SampledWavelengths}; -use crate::utils::ptr::DevicePtr; +use crate::utils::Ptr; use crate::{Float, PI}; #[repr(C)] @@ -15,7 +15,7 @@ use crate::{Float, PI}; pub struct PointLight { pub base: LightBase, pub scale: Float, - pub i: DevicePtr, + pub i: Ptr, } impl LightTrait for PointLight { diff --git a/shared/src/lights/projection.rs b/shared/src/lights/projection.rs index 2542c77..9b0b067 100644 --- a/shared/src/lights/projection.rs +++ b/shared/src/lights/projection.rs @@ -11,10 +11,9 @@ use crate::core::medium::MediumInterface; use crate::core::spectrum::SpectrumTrait; use crate::spectra::{SampledSpectrum, SampledWavelengths}; use crate::utils::math::{radians, square}; -use crate::utils::ptr::DevicePtr; use crate::{ spectra::{RGBColorSpace, RGBIlluminantSpectrum}, - utils::{Transform, sampling::DevicePiecewiseConstant2D}, + utils::{Ptr, Transform, sampling::DevicePiecewiseConstant2D}, }; #[repr(C)] @@ -27,9 +26,9 @@ pub struct ProjectionLight { pub screen_from_light: Transform, pub light_from_screen: Transform, pub a: Float, - pub image: DevicePtr, - pub distrib: DevicePtr, - pub image_color_space: DevicePtr, + pub image: Ptr, + pub distrib: Ptr, + pub image_color_space: Ptr, } impl ProjectionLight { diff --git a/shared/src/lights/spot.rs b/shared/src/lights/spot.rs index fbd8c92..f5ee84f 100644 --- a/shared/src/lights/spot.rs +++ b/shared/src/lights/spot.rs @@ -5,14 +5,14 @@ use crate::core::interaction::{Interaction, InteractionBase, InteractionTrait, S use crate::core::light::{LightBase, LightBounds, LightLiSample, LightSampleContext, LightTrait}; use crate::core::spectrum::SpectrumTrait; use crate::spectra::{DenselySampledSpectrum, SampledSpectrum, SampledWavelengths}; -use crate::utils::DevicePtr; +use crate::utils::Ptr; use crate::{Float, PI}; #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct SpotLight { pub base: LightBase, - pub iemit: DevicePtr, + pub iemit: Ptr, pub scale: Float, pub cos_falloff_start: Float, pub cos_falloff_end: Float, diff --git a/shared/src/materials/coated.rs b/shared/src/materials/coated.rs index 2433276..36b95f5 100644 --- a/shared/src/materials/coated.rs +++ b/shared/src/materials/coated.rs @@ -10,21 +10,21 @@ use crate::core::scattering::TrowbridgeReitzDistribution; use crate::core::spectrum::{Spectrum, SpectrumTrait}; use crate::core::texture::{GPUFloatTexture, GPUSpectrumTexture, TextureEvaluator}; use crate::spectra::{SampledSpectrum, SampledWavelengths}; -use crate::utils::DevicePtr; +use crate::utils::Ptr; use crate::utils::math::clamp; #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct CoatedDiffuseMaterial { - pub normal_map: DevicePtr, - pub displacement: DevicePtr, - pub reflectance: DevicePtr, - pub albedo: DevicePtr, - pub u_roughness: DevicePtr, - pub v_roughness: DevicePtr, - pub thickness: DevicePtr, - pub g: DevicePtr, - pub eta: DevicePtr, + pub normal_map: Ptr, + pub displacement: Ptr, + pub reflectance: Ptr, + pub albedo: Ptr, + pub u_roughness: Ptr, + pub v_roughness: Ptr, + pub thickness: Ptr, + pub g: Ptr, + pub eta: Ptr, pub remap_roughness: bool, pub max_depth: usize, pub n_samples: usize, @@ -48,15 +48,15 @@ impl CoatedDiffuseMaterial { n_samples: usize, ) -> Self { Self { - displacement: DevicePtr::from(displacement), - normal_map: DevicePtr::from(normal_map), - reflectance: DevicePtr::from(reflectance), - albedo: DevicePtr::from(albedo), - u_roughness: DevicePtr::from(u_roughness), - v_roughness: DevicePtr::from(v_roughness), - thickness: DevicePtr::from(thickness), - g: DevicePtr::from(g), - eta: DevicePtr::from(eta), + displacement: Ptr::from(displacement), + normal_map: Ptr::from(normal_map), + reflectance: Ptr::from(reflectance), + albedo: Ptr::from(albedo), + u_roughness: Ptr::from(u_roughness), + v_roughness: Ptr::from(v_roughness), + thickness: Ptr::from(thickness), + g: Ptr::from(g), + eta: Ptr::from(eta), remap_roughness, max_depth, n_samples, @@ -115,7 +115,7 @@ impl MaterialTrait for CoatedDiffuseMaterial { self.n_samples, )); - BSDF::new(ctx.ns, ctx.dpdus, DevicePtr::from(&bxdf)) + BSDF::new(ctx.ns, ctx.dpdus, Ptr::from(&bxdf)) } fn get_bssrdf( @@ -138,7 +138,7 @@ impl MaterialTrait for CoatedDiffuseMaterial { Some(&*self.normal_map) } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { self.displacement } @@ -150,19 +150,19 @@ impl MaterialTrait for CoatedDiffuseMaterial { #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct CoatedConductorMaterial { - normal_map: DevicePtr, - displacement: DevicePtr, - interface_uroughness: DevicePtr, - interface_vroughness: DevicePtr, - thickness: DevicePtr, - interface_eta: DevicePtr, - g: DevicePtr, - albedo: DevicePtr, - conductor_uroughness: DevicePtr, - conductor_vroughness: DevicePtr, - conductor_eta: DevicePtr, - k: DevicePtr, - reflectance: DevicePtr, + normal_map: Ptr, + displacement: Ptr, + interface_uroughness: Ptr, + interface_vroughness: Ptr, + thickness: Ptr, + interface_eta: Ptr, + g: Ptr, + albedo: Ptr, + conductor_uroughness: Ptr, + conductor_vroughness: Ptr, + conductor_eta: Ptr, + k: Ptr, + reflectance: Ptr, remap_roughness: bool, max_depth: u32, n_samples: u32, @@ -190,19 +190,19 @@ impl CoatedConductorMaterial { n_samples: u32, ) -> Self { Self { - displacement: DevicePtr::from(displacement), - normal_map: DevicePtr::from(normal_map), - interface_uroughness: DevicePtr::from(interface_uroughness), - interface_vroughness: DevicePtr::from(interface_vroughness), - thickness: DevicePtr::from(thickness), - interface_eta: DevicePtr::from(interface_eta), - g: DevicePtr::from(g), - albedo: DevicePtr::from(albedo), - conductor_uroughness: DevicePtr::from(conductor_uroughness), - conductor_vroughness: DevicePtr::from(conductor_vroughness), - conductor_eta: DevicePtr::from(conductor_eta), - k: DevicePtr::from(k), - reflectance: DevicePtr::from(reflectance), + displacement: Ptr::from(displacement), + normal_map: Ptr::from(normal_map), + interface_uroughness: Ptr::from(interface_uroughness), + interface_vroughness: Ptr::from(interface_vroughness), + thickness: Ptr::from(thickness), + interface_eta: Ptr::from(interface_eta), + g: Ptr::from(g), + albedo: Ptr::from(albedo), + conductor_uroughness: Ptr::from(conductor_uroughness), + conductor_vroughness: Ptr::from(conductor_vroughness), + conductor_eta: Ptr::from(conductor_eta), + k: Ptr::from(k), + reflectance: Ptr::from(reflectance), remap_roughness, max_depth, n_samples, @@ -285,7 +285,7 @@ impl MaterialTrait for CoatedConductorMaterial { self.max_depth as usize, self.n_samples as usize, )); - BSDF::new(ctx.ns, ctx.dpdus, DevicePtr::from(&bxdf)) + BSDF::new(ctx.ns, ctx.dpdus, Ptr::from(&bxdf)) } fn get_bssrdf( @@ -330,7 +330,7 @@ impl MaterialTrait for CoatedConductorMaterial { Some(&*self.normal_map) } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { self.displacement } diff --git a/shared/src/materials/conductor.rs b/shared/src/materials/conductor.rs index 3454cb3..2353384 100644 --- a/shared/src/materials/conductor.rs +++ b/shared/src/materials/conductor.rs @@ -10,20 +10,20 @@ use crate::core::scattering::TrowbridgeReitzDistribution; use crate::core::spectrum::{Spectrum, SpectrumTrait}; use crate::core::texture::{GPUFloatTexture, GPUSpectrumTexture, TextureEvaluator}; use crate::spectra::{SampledSpectrum, SampledWavelengths}; -use crate::utils::DevicePtr; +use crate::utils::Ptr; use crate::utils::math::clamp; #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct ConductorMaterial { - pub displacement: DevicePtr, - pub eta: DevicePtr, - pub k: DevicePtr, - pub reflectance: DevicePtr, - pub u_roughness: DevicePtr, - pub v_roughness: DevicePtr, + pub displacement: Ptr, + pub eta: Ptr, + pub k: Ptr, + pub reflectance: Ptr, + pub u_roughness: Ptr, + pub v_roughness: Ptr, pub remap_roughness: bool, - pub normal_map: DevicePtr, + pub normal_map: Ptr, } impl MaterialTrait for ConductorMaterial { @@ -54,7 +54,7 @@ impl MaterialTrait for ConductorMaterial { todo!() } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { todo!() } diff --git a/shared/src/spectra/simple.rs b/shared/src/spectra/simple.rs index 4d2fcd6..9738f01 100644 --- a/shared/src/spectra/simple.rs +++ b/shared/src/spectra/simple.rs @@ -1,9 +1,10 @@ use super::cie::*; use super::sampled::{LAMBDA_MAX, LAMBDA_MIN}; +use crate::Float; use crate::core::spectrum::{Spectrum, SpectrumTrait}; use crate::spectra::{N_SPECTRUM_SAMPLES, SampledSpectrum, SampledWavelengths}; +use crate::utils::find_interval; use crate::utils::ptr::DevicePtr; -use crate::{Float, find_interval}; use core::slice; use std::hash::{Hash, Hasher}; use std::sync::LazyLock; diff --git a/shared/src/utils/containers.rs b/shared/src/utils/containers.rs index 4125629..0746557 100644 --- a/shared/src/utils/containers.rs +++ b/shared/src/utils/containers.rs @@ -25,7 +25,7 @@ impl Interpolatable for T where pub struct Array2D { pub values: *mut T, pub extent: Bounds2i, - pub x_stride: i32, + pub stride: i32, } unsafe impl Send for Array2D {} @@ -51,7 +51,7 @@ impl Array2D { fn offset(&self, p: Point2i) -> isize { let ox = p.x() - self.extent.p_min.x(); let oy = p.y() - self.extent.p_min.y(); - (ox + oy * self.x_stride) as isize + (ox + oy * self.stride) as isize } #[inline] diff --git a/shared/src/utils/mod.rs b/shared/src/utils/mod.rs index 0fdcf82..e412715 100644 --- a/shared/src/utils/mod.rs +++ b/shared/src/utils/mod.rs @@ -18,6 +18,30 @@ pub mod transform; pub use ptr::{DevicePtr, Ptr}; pub use transform::{AnimatedTransform, Transform, TransformGeneric}; +#[inline] +pub fn find_interval(sz: u32, pred: F) -> u32 +where + F: Fn(u32) -> bool, +{ + let mut first = 0; + let mut len = sz; + + while len > 0 { + let half = len >> 1; + let middle = first + half; + + if pred(middle) { + first = middle + 1; + len -= half + 1; + } else { + len = half; + } + } + + let ret = (first as i32 - 1).max(0) as u32; + ret.min(sz.saturating_sub(2)) +} + #[inline] pub fn partition_slice(data: &mut [T], predicate: F) -> usize where diff --git a/shared/src/utils/ptr.rs b/shared/src/utils/ptr.rs index 56359bc..a02e3b8 100644 --- a/shared/src/utils/ptr.rs +++ b/shared/src/utils/ptr.rs @@ -1,11 +1,10 @@ use core::marker::PhantomData; use core::ops::Index; -#[repr(C)] +#[repr(transparent)] #[derive(Debug)] pub struct Ptr { - pub offset: u32, - pub _marker: PhantomData, + ptr: *const T, } impl Clone for Ptr { @@ -15,155 +14,123 @@ impl Clone for Ptr { } impl Copy for Ptr {} +// Ptr is just a pointer - Send/Sync depends on T +unsafe impl Send for Ptr {} +unsafe impl Sync for Ptr {} + impl Ptr { - pub const NULL_OFFSET: u32 = 0xFFFFFFFF; - pub fn null() -> Self { + pub const fn null() -> Self { Self { - offset: Self::NULL_OFFSET, - _marker: PhantomData, + ptr: std::ptr::null(), } } - pub fn is_null(&self) -> bool { - self.offset == Self::NULL_OFFSET + pub const fn is_null(self) -> bool { + self.ptr.is_null() + } + + pub fn from_raw(ptr: *const T) -> Self { + Self { ptr } + } + + pub fn as_raw(self) -> *const T { + self.ptr } #[inline(always)] - pub unsafe fn deref<'a>(&self, base: *const u8) -> &'a T { + pub unsafe fn as_ref<'a>(self) -> &'a T { + debug_assert!(!self.is_null(), "null Ptr dereference"); + unsafe { &*self.ptr } + } + + /// Get as Option - safe for optional fields + #[inline(always)] + pub fn get<'a>(self) -> Option<&'a T> { if self.is_null() { - panic!("Null pointer dereference"); + None + } else { + Some(unsafe { &*self.ptr }) } - unsafe { - let ptr = base.add(self.offset as usize); - &*(ptr as *const T) + } + + #[inline(always)] + pub unsafe fn at<'a>(self, index: usize) -> &'a T { + debug_assert!(!self.is_null(), "null Ptr array access"); + unsafe { &*self.ptr.add(index) } + } + + /// Get element at index, returns None if ptr is null + #[inline(always)] + pub fn get_at<'a>(self, index: usize) -> Option<&'a T> { + if self.is_null() { + None + } else { + Some(unsafe { &*self.ptr.add(index) }) + } + } + + /// Offset the pointer, returning a new Ptr + #[inline(always)] + pub unsafe fn add(self, count: usize) -> Self { + Self { + ptr: unsafe { self.ptr.add(count) }, + } + } + + /// Convert to slice (requires knowing the length) + #[inline(always)] + pub unsafe fn as_slice<'a>(self, len: usize) -> &'a [T] { + debug_assert!(!self.is_null() || len == 0, "null Ptr to non-empty slice"); + if len == 0 { + &[] + } else { + unsafe { std::slice::from_raw_parts(self.ptr, len) } + } + } + + /// Convert to slice, returns None if null + #[inline(always)] + pub fn get_slice<'a>(self, len: usize) -> Option<&'a [T]> { + if self.is_null() { + if len == 0 { Some(&[]) } else { None } + } else { + Some(unsafe { std::slice::from_raw_parts(self.ptr, len) }) } } } -#[repr(transparent)] -#[derive(Debug, Clone, Copy)] -pub struct DevicePtr(pub *const T); - -impl Default for DevicePtr { - fn default() -> Self { - Self(core::ptr::null()) - } -} - -impl PartialEq for DevicePtr { - #[inline(always)] - fn eq(&self, other: &Self) -> bool { - self.0 == other.0 - } -} - -impl Eq for DevicePtr {} - -impl DevicePtr { - pub fn null() -> Self { - Self::default() - } - - #[inline(always)] - pub fn is_null(&self) -> bool { - self.0.is_null() - } - - #[inline(always)] - pub fn as_ref(&self) -> Option<&T> { - unsafe { self.0.as_ref() } - } - - /// UNSTABLE: Casts the const pointer to mutable and returns a mutable reference. - /// THIS IS VERY DANGEROUS - /// The underlying data is not currently borrowed or accessed by any other thread/kernel. - /// The memory is actually writable. - /// No other mutable references exist to this data. - #[inline(always)] - pub unsafe fn as_mut(&mut self) -> &mut T { - debug_assert!(!self.is_null()); - unsafe { &mut *(self.0 as *mut T) } - } -} - -unsafe impl Send for DevicePtr {} -unsafe impl Sync for DevicePtr {} - -impl std::ops::Deref for DevicePtr { +impl std::ops::Deref for Ptr { type Target = T; #[inline(always)] - fn deref(&self) -> &Self::Target { - unsafe { &*self.0 } + fn deref(&self) -> &T { + debug_assert!(!self.is_null(), "Null Ptr dereference"); + unsafe { &*self.ptr } } } -impl From<&T> for DevicePtr { +impl Default for Ptr { + fn default() -> Self { + Self::null() + } +} + +impl From<&T> for Ptr { fn from(r: &T) -> Self { - Self(r as *const T) + Self { ptr: r as *const T } } } -impl From<&mut T> for DevicePtr { - fn from(r: &mut T) -> Self { - Self(r as *const T) - } -} - -impl Index for DevicePtr { - type Output = T; - - #[inline(always)] - fn index(&self, index: usize) -> &Self::Output { - // There is no bounds checking because we dont know the length. - // It is host responsbility to check bounds - unsafe { &*self.0.add(index) } - } -} - -impl Index for DevicePtr { - type Output = T; - - #[inline(always)] - fn index(&self, index: u32) -> &Self::Output { - unsafe { &*self.0.add(index as usize) } - } -} - -#[repr(C)] -#[derive(Clone, Copy, Debug)] -pub struct Slice { - pub ptr: DevicePtr, - pub len: u32, - pub _marker: PhantomData, -} - -unsafe impl Send for Slice {} -unsafe impl Sync for Slice {} - -impl Slice { - pub fn new(ptr: DevicePtr, len: u32) -> Self { +impl From<&[T]> for Ptr { + fn from(slice: &[T]) -> Self { Self { - ptr, - len, - _marker: PhantomData, + ptr: slice.as_ptr(), } } - - pub fn as_ptr(&self) -> *const T { - self.ptr.0 - } - - pub fn is_empty(&self) -> bool { - self.len == 0 - } } -impl Index for Slice { - type Output = T; - - #[inline(always)] - fn index(&self, index: usize) -> &Self::Output { - unsafe { &*self.ptr.0.add(index) } +impl From<*const T> for Ptr { + fn from(ptr: *const T) -> Self { + Self { ptr } } } diff --git a/shared/src/utils/sampling.rs b/shared/src/utils/sampling.rs index 3389cf1..94bac16 100644 --- a/shared/src/utils/sampling.rs +++ b/shared/src/utils/sampling.rs @@ -4,15 +4,14 @@ use crate::core::geometry::{ }; use crate::core::pbrt::{RARE_EVENT_CONDITION_MET, RARE_EVENT_TOTAL_CALLS}; use crate::utils::containers::Array2D; +use crate::utils::find_interval; use crate::utils::math::{ catmull_rom_weights, clamp, difference_of_products, evaluate_polynomial, lerp, logistic, newton_bisection, safe_sqrt, square, sum_of_products, }; -use crate::utils::ptr::DevicePtr; +use crate::utils::ptr::Ptr; use crate::utils::rng::Rng; -use crate::{ - Float, INV_2_PI, INV_4_PI, INV_PI, ONE_MINUS_EPSILON, PI, PI_OVER_2, PI_OVER_4, find_interval, -}; +use crate::{Float, INV_2_PI, INV_4_PI, INV_PI, ONE_MINUS_EPSILON, PI, PI_OVER_2, PI_OVER_4}; use num_traits::Num; pub fn linear_pdf(x: T, a: T, b: T) -> T @@ -703,8 +702,8 @@ pub struct PLSample { #[repr(C)] #[derive(Debug, Copy, Clone)] pub struct DevicePiecewiseConstant1D { - pub func: DevicePtr, - pub cdf: DevicePtr, + pub func: Ptr, + pub cdf: Ptr, pub min: Float, pub max: Float, pub n: u32, @@ -723,39 +722,48 @@ impl DevicePiecewiseConstant1D { self.n } - pub fn sample(&self, u: Float) -> (Float, Float, u32) { - let o = find_interval(self.size(), |idx| self.cdf[idx] <= u) as usize; - let mut du = u - self.cdf[o]; - if self.cdf[o + 1] - self.cdf[o] > 0. { - du /= self.cdf[o + 1] - self.cdf[o]; - } - debug_assert!(!du.is_nan()); - let value = lerp((o as Float + du) / self.size() as Float, self.min, self.max); - let pdf_val = if self.func_integral > 0. { - self.func[o] / self.func_integral + pub fn sample(&self, u: Float) -> (Float, Float, usize) { + // Find offset via binary search on CDF + let offset = self.find_interval(u); + + let cdf_offset = unsafe { *self.cdf.add(offset) }; + let cdf_next = unsafe { *self.cdf.add(offset + 1) }; + + let du = if cdf_next - cdf_offset > 0.0 { + (u - cdf_offset) / (cdf_next - cdf_offset) } else { - 0. + 0.0 }; - (value, pdf_val, o as u32) + + let delta = (self.max - self.min) / self.n as Float; + let x = self.min + (offset as Float + du) * delta; + + let pdf = if self.func_integral > 0.0 { + (unsafe { *self.func.add(offset) }) / self.func_integral + } else { + 0.0 + }; + + (x, pdf, offset) } } #[repr(C)] #[derive(Debug, Copy, Clone)] pub struct DevicePiecewiseConstant2D { - pub domain: Bounds2f, - pub p_marginal: DevicePiecewiseConstant1D, - pub n_conditionals: usize, - pub p_conditional_v: DevicePtr, + pub conditional: *const DevicePiecewiseConstant1D, // Array of n_v conditionals + pub marginal: DevicePiecewiseConstant1D, + pub n_u: u32, + pub n_v: u32, } impl DevicePiecewiseConstant2D { - pub fn resolution(&self) -> Point2i { - Point2i::new( - self.p_conditional_v[0u32].size() as i32, - self.p_conditional_v[1u32].size() as i32, - ) - } + // pub fn resolution(&self) -> Point2i { + // Point2i::new( + // self.p_conditional_v[0u32].size() as i32, + // self.p_conditional_v[1u32].size() as i32, + // ) + // } pub fn integral(&self) -> f32 { self.p_marginal.integral() @@ -769,20 +777,20 @@ impl DevicePiecewiseConstant2D { (Point2f::new(d0, d1), pdf, offset) } - pub fn pdf(&self, p: Point2f) -> f32 { - let p_offset = self.domain.offset(&p); - let nu = self.p_conditional_v[0u32].size(); - let nv = self.p_marginal.size(); + pub fn pdf(&self, p: Point2f) -> Float { + // Find which row + let delta_v = 1.0 / self.n_v as Float; + let v_offset = ((p.y() * self.n_v as Float) as usize).min(self.n_v as usize - 1); - let iu = (p_offset.x() * nu as f32).clamp(0.0, nu as f32 - 1.0) as usize; - let iv = (p_offset.y() * nv as f32).clamp(0.0, nv as f32 - 1.0) as usize; + let conditional = unsafe { &*self.conditional.add(v_offset) }; - let integral = self.p_marginal.integral(); - if integral == 0.0 { - 0.0 - } else { - self.p_conditional_v[iv].func[iu] / integral - } + // Find which column + let delta_u = 1.0 / self.n_u as Float; + let u_offset = ((p.x() * self.n_u as Float) as usize).min(self.n_u as usize - 1); + + let func_val = unsafe { *conditional.func.add(u_offset) }; + + func_val / self.marginal.func_integral } } @@ -1056,10 +1064,10 @@ pub struct PiecewiseLinear2D { pub inv_patch_size: Vector2f, pub param_size: [u32; N], pub param_strides: [u32; N], - pub param_values: [DevicePtr; N], - pub data: DevicePtr, - pub marginal_cdf: DevicePtr, - pub conditional_cdf: DevicePtr, + pub param_values: [Ptr; N], + pub data: Ptr, + pub marginal_cdf: Ptr, + pub conditional_cdf: Ptr, } impl PiecewiseLinear2D { @@ -1311,7 +1319,7 @@ impl PiecewiseLinear2D { fn lookup( &self, - data: DevicePtr, + data: Ptr, i0: u32, size: u32, param_weight: &[(Float, Float); N], diff --git a/src/core/aggregates.rs b/src/core/aggregates.rs index 4178ba5..116742a 100644 --- a/src/core/aggregates.rs +++ b/src/core/aggregates.rs @@ -1,10 +1,10 @@ +use crate::Float; use crate::core::geometry::{Bounds3f, Point3f, Ray, Vector3f}; -use crate::core::pbrt::{Float, find_interval}; use crate::core::primitive::PrimitiveTrait; use crate::core::shape::ShapeIntersection; use crate::utils::math::encode_morton_3; use crate::utils::math::next_float_down; -use crate::utils::partition_slice; +use crate::utils::{find_interval, partition_slice}; use rayon::prelude::*; use std::cmp::Ordering; use std::sync::Arc; @@ -13,7 +13,7 @@ use std::sync::atomic::{AtomicUsize, Ordering as AtomicOrdering}; #[repr(C)] #[derive(Debug, Clone, Copy, PartialEq, Eq)] pub enum SplitMethod { - SAH, + AH, Hlbvh, Middle, EqualCounts, @@ -345,7 +345,6 @@ impl BVHAggregate { 4, ); - // Add thread-local count to global atomic total_nodes.fetch_add(nodes_created, AtomicOrdering::Relaxed); root diff --git a/src/core/camera.rs b/src/core/camera.rs index 23b196f..8d6bd85 100644 --- a/src/core/camera.rs +++ b/src/core/camera.rs @@ -1,5 +1,5 @@ use crate::core::image::ImageMetadata; -use crate::utils::{FileLoc, ParameterDictionary}; +use crate::utils::{Arena, FileLoc, ParameterDictionary}; use shared::Float; use shared::cameras::*; use shared::core::camera::{Camera, CameraBase, CameraTrait, CameraTransform}; @@ -99,6 +99,7 @@ impl CameraFactory for Camera { medium: Medium, film: Arc, loc: &FileLoc, + arena: &mut Arena, ) -> Result { match name { "perspective" => { @@ -143,6 +144,7 @@ impl CameraFactory for Camera { let fov = params.get_one_float("fov", 90.); let camera = PerspectiveCamera::new(base, fov, screen, lens_radius, focal_distance); + arena.alloc(camera); Ok(Camera::Perspective(camera)) } @@ -187,6 +189,7 @@ impl CameraFactory for Camera { } } let camera = OrthographicCamera::new(base, screen, lens_radius, focal_distance); + arena.alloc(camera); Ok(Camera::Orthographic(camera)) } "realistic" => { @@ -367,6 +370,7 @@ impl CameraFactory for Camera { aperture_image, ); + arena.alloc(camera); Ok(Camera::Realistic(camera)) } "spherical" => { @@ -423,6 +427,7 @@ impl CameraFactory for Camera { let camera = SphericalCamera { mapping, base }; + arena.alloc(camera); Ok(Camera::Spherical(camera)) } _ => Err(format!("Camera type '{}' unknown at {}", name, loc)), diff --git a/src/core/image/io.rs b/src/core/image/io.rs index ece3215..921374c 100644 --- a/src/core/image/io.rs +++ b/src/core/image/io.rs @@ -206,14 +206,14 @@ fn read_generic(path: &Path, encoding: Option) -> Result DeviceImage { + DynamicImage::ImageRgb32F(buf) => Image { format: PixelFormat::F32, resolution: res, channel_names: vec!["R".into(), "G".into(), "B".into()], encoding: LINEAR, pixels: PixelData::F32(buf.into_raw()), }, - DynamicImage::ImageRgba32F(buf) => DeviceImage { + DynamicImage::ImageRgba32F(buf) => Image { format: PixelFormat::F32, resolution: res, channel_names: vec!["R".into(), "G".into(), "B".into(), "A".into()], @@ -224,7 +224,7 @@ fn read_generic(path: &Path, encoding: Option) -> Result) -> Result Result { let w = image.layer_data.size.width() as i32; let h = image.layer_data.size.height() as i32; - let image = DeviceImage { + let image = Image { format: PixelFormat::F32, resolution: Point2i::new(w, h), channel_names: vec!["R".into(), "G".into(), "B".into(), "A".into()], @@ -355,7 +355,7 @@ fn read_pfm(path: &Path) -> Result { vec!["R".into(), "G".into(), "B".into()] }; - let image = DeviceImage { + let image = Image { format: PixelFormat::F32, resolution: Point2i::new(w, h), channel_names: names, diff --git a/src/core/image/mod.rs b/src/core/image/mod.rs index 7ae6c65..b7d4e22 100644 --- a/src/core/image/mod.rs +++ b/src/core/image/mod.rs @@ -71,30 +71,41 @@ impl DerefMut for ImageChannelValues { #[derive(Debug, Clone)] pub enum PixelStorage { - U8(Vec), - F16(Vec), - F32(Vec), + U8(Box<[u8]>), + F16(Box<[f16]>), + F32(Box<[f32]>), +} + +impl PixelStorage { + pub fn as_pixels(&self) -> Pixels { + match self { + PixelStorage::U8(data) => Pixels::U8(data.as_ptr()), + PixelStorage::F16(data) => Pixels::F16(data.as_ptr() as *const u16), + PixelStorage::F32(data) => Pixels::F32(data.as_ptr()), + } + } + + pub fn format(&self) -> PixelFormat { + match self { + PixelStorage::U8(_) => PixelFormat::U8, + PixelStorage::F16(_) => PixelFormat::F16, + PixelStorage::F32(_) => PixelFormat::F32, + } + } + + pub fn len(&self) -> usize { + match self { + PixelStorage::U8(d) => d.len(), + PixelStorage::F16(d) => d.len(), + PixelStorage::F32(d) => d.len(), + } + } } pub struct Image { - pub base: ImageBase, - pub pixels: PixelStorage, - pub channel_names: Vec, -} - -impl ImageAccess for Image { - fn get_channel_with_wrap(&self, p: Point2i, c: i32, wrap_mode: WrapMode2D) -> Float { - todo!() - } - fn get_channel(&self, p: Point2i, c: i32) -> Float { - todo!() - } - fn lookup_nearest_channel_with_wrap(&self, p: Point2f, c: i32, wrap_mode: WrapMode2D) -> Float { - todo!() - } - fn lookup_nearest_channel(&self, p: Point2f, c: i32) -> Float { - todo!() - } + storage: PixelStorage, + channel_names: Vec, + device: DeviceImage, } #[derive(Debug, Clone)] @@ -104,52 +115,46 @@ pub struct ImageAndMetadata { } impl Image { - pub fn resolution(&self) -> Point2i { - self.base.resolution - } + // Constructors + fn from_storage( + storage: PixelStorage, + resolution: Point2i, + channel_names: Vec, + encoding: ColorEncoding, + ) -> Self { + let n_channels = channel_names.len() as i32; + let expected = (resolution.x() * resolution.y()) as usize * n_channels as usize; + assert_eq!(storage.len(), expected, "Pixel data size mismatch"); - fn n_channels(&self) -> i32 { - self.base.n_channels - } - - pub fn new_empty() -> Self { - Self { - channel_names: Vec::new(), - pixels: PixelStorage::U8(Vec::new()), + let device = DeviceImage { base: ImageBase { - format: PixelFormat::U256, - resolution: Point2i::new(0, 0), - n_channels: 0, - encoding: ColorEncoding::default(), + format: storage.format(), + encoding, + resolution, + n_channels, }, + pixels: storage.as_pixels(), + }; + + Self { + storage, + channel_names, + device, } } - fn from_vector( - format: PixelFormat, + pub fn from_u8( + data: Vec, resolution: Point2i, channel_names: Vec, encoding: ColorEncoding, ) -> Self { - let n_channels = channel_names.len() as i32; - let (format, pixels) = match &storage { - PixelStorage::U8(vec) => (PixelFormat::U8, ImagePixels::U8(vec.as_ptr())), - PixelStorage::F16(vec) => (PixelFormat::F16, ImagePixels::F16(vec.as_ptr())), - PixelStorage::F32(vec) => (PixelFormat::F32, ImagePixels::F32(vec.as_ptr())), - }; - - let base = ImageBase { - format, + Self::from_storage( + PixelStorage::U8(data.into_boxed_slice()), resolution, - n_channels, + channel_names, encoding, - }; - - Self { - base, - pixels, - channel_names, - } + ) } pub fn from_u8( @@ -158,133 +163,30 @@ impl Image { channel_names: Vec, encoding: ColorEncoding, ) -> Self { - let n_channels = channel_names.len() as i32; - let expected_len = (resolution.x * resolution.y) as usize * n_channels as usize; - if data.len() != expected_len { - panic!( - "ImageBuffer::from_u8: Data length {} does not match resolution {:?} * channels {}", - data.len(), - resolution, - n_channels - ); - } - - let storage = PixelStorage::U8(data); - - let ptr = match &storage { - PixelStorage::U8(v) => v.as_ptr(), - _ => unreachable!(), - }; - - let base = ImageBase { - format: PixelFormat::U256, + Self::from_storage( + PixelStorage::U8(data.into_boxed_slice()), resolution, - n_channels, - encoding: encoding.clone(), - }; - - Self { - base, channel_names, - pixels: ptr, - } - } - - pub fn from_u8( - data: Vec, - resolution: Point2i, - channel_names: Vec, - encoding: ColorEncoding, - ) -> Self { - let n_channels = channel_names.len() as i32; - let expected_len = (resolution.x * resolution.y) as usize * n_channels as usize; - - if data.len() != expected_len { - panic!( - "ImageBuffer::from_u8: Data length {} does not match resolution {:?} * channels {}", - data.len(), - resolution, - n_channels - ); - } - - let storage = PixelStorage::U8(data); - - let ptr = match &storage { - PixelStorage::U8(v) => v.as_ptr(), - _ => unreachable!(), - }; - - let base = ImageBase { - format: PixelFormat::U256, - resolution, - n_channels, - encoding: encoding.clone(), - }; - - Self { - base, - channel_names, - pixels: ptr, - } + encoding, + ) } pub fn from_f16(data: Vec, resolution: Point2i, channel_names: Vec) -> Self { - let n_channels = channel_names.len() as i32; - let expected_len = (resolution.x * resolution.y) as usize * n_channels as usize; - - if data.len() != expected_len { - panic!("ImageBuffer::from_f16: Data length mismatch"); - } - - let storage = PixelStorage::F16(data); - - let ptr = match &storage { - PixelStorage::F16(v) => v.as_ptr() as *const u16, - _ => unreachable!(), - }; - - let base = ImageBase { - format: PixelFormat::Half, + Self::from_storage( + PixelStorage::F16(data.into_boxed_slice()), resolution, - n_channels, - encoding: ColorEncoding::default(), - }; - - Self { - base, channel_names, - pixels: ptr, - } + ColorEncoding::Linear, + ) } pub fn from_f32(data: Vec, resolution: Point2i, channel_names: Vec) -> Self { - let n_channels = channel_names.len() as i32; - let expected_len = (resolution.x * resolution.y) as usize * n_channels as usize; - - if data.len() != expected_len { - panic!("ImageBuffer::from_f32: Data length mismatch"); - } - - let storage = PixelStorage::F32(data); - - let ptr = match &storage { - PixelStorage::F32(v) => v.as_ptr(), - _ => unreachable!(), - }; - - let base = ImageBase { - format: PixelFormat::Float, + Self::from_storage( + PixelStorage::F32(data.into_boxed_slice()), resolution, - n_channels, - encoding: ColorEncoding::default(), - }; - - Self { - base, channel_names, - pixels: ptr, - } + ColorEncoding::Linear, + ) } pub fn new( @@ -303,7 +205,24 @@ impl Image { PixelFormat::F32 => PixelStorage::F32(vec![0.0; pixel_count]), }; - Self::from_vector(storage, resolution, owned_names, encoding) + Self::from_storage(storage, resolution, owned_names, encoding) + } + + // Access + pub fn device_image(&self) -> &DeviceImage { + &self.device + } + + pub fn resolution(&self) -> Point2i { + self.base.resolution + } + + fn n_channels(&self) -> i32 { + self.base.n_channels + } + + pub fn format(&self) -> PixelFormat { + self.device.base.format } pub fn channel_names(&self) -> Vec<&str> { @@ -314,104 +233,62 @@ impl Image { self.view.encoding } - pub fn set_channel(&self, p: Point2i, c: i32, mut value: Float) { - if value.is_nan() { - value = 0.0; - } - - if !self.view.resolution.inside_exclusive(p) { - return; - } - - let offset = self.view.pixel_offset(p) + c as usize; - match &mut self._storage { - PixelStorage::U8(data) => { - if !self.view.encoding.is_null() { - data[offset] = self.view.encoding.from_linear_scalar(value); - } else { - let val = (value * 255.0 + 0.5).clamp(0.0, 255.0); - data[offset] = val as u8; - } - } - PixelStorage::F16(data) => { - data[offset] = f16_to_f32(value); - } - PixelStorage::F32(data) => { - data[offset] = value; - } - } + fn pixel_offset(&self, p: Point2i) -> usize { + let width = self.resolution().x() as usize; + let idx = p.y() as usize * width + p.x() as usize; + idx * self.n_channels() as usize } - pub fn get_channels_with_wrap(&self, p: Point2i, wrap_mode: WrapMode2D) -> ImageChannelValues { - let mut pp = p; - if !self.view.remap_pixel_coords(&mut pp, wrap_mode) { - return ImageChannelValues(smallvec![0.0; desc.offset.len()]); + // Read + pub fn get_channel(&self, p: Point2i, c: i32) -> Float { + self.get_channel_with_wrap(p, c, WrapMode::Clamp.into()) + } + + pub fn get_channel_with_wrap(&self, mut p: Point2i, c: i32, wrap_mode: WrapMode2D) -> Float { + if !self.device.base.remap_pixel_coords(&mut p, wrap_mode) { + return 0.0; } - let pixel_offset = self.view.pixel_offset(pp); - let mut values: SmallVec<[Float; 4]> = SmallVec::with_capacity(desc.offset.len()); - match &self.pixels { - PixelData::U8(data) => { - for i in 0..self.view.n_channels() { - let raw_val = data[pixel_offset + i]; - let linear = self.view.encoding.to_linear_scalar(raw_val); - values.push(linear); - } - } - PixelData::F16(data) => { - for i in 0..self.view.n_channels() { - let raw_val = data[pixel_offset]; - values.push(f16_to_f32(raw_val)); - } - } - PixelData::F32(data) => { - for i in 0..self.view.n_channels() { - let val = data[pixel_offset + i]; - values.push(val); - } - } - } + let offset = self.pixel_offset(p) + c as usize; - ImageChannelValues(values) + match &self.storage { + PixelStorage::U8(data) => self.device.base.encoding.to_linear_scalar(data[offset]), + PixelStorage::F16(data) => data[offset].to_f32(), + PixelStorage::F32(data) => data[offset], + } } pub fn get_channels(&self, p: Point2i) -> ImageChannelValues { self.get_channels_with_wrap(p, WrapMode::Clamp.into()) } - pub fn get_channels_with_desc( + pub fn get_channels_with_wrap( &self, - p: Point2i, - desc: &ImageChannelDesc, - wrap: WrapMode2D, + mut p: Point2i, + wrap_mode: WrapMode2D, ) -> ImageChannelValues { - let mut pp = p; - if !self.view.remap_pixel_coords(&mut pp, wrap) { - return ImageChannelValues(smallvec![0.0; desc.offset.len()]); + if !self.device.base.remap_pixel_coords(&mut p, wrap_mode) { + return ImageChannelValues(smallvec![0.0; self.n_channels() as usize]); } - let pixel_offset = self.view.pixel_offset(pp); + let offset = self.pixel_offset(p); + let nc = self.n_channels() as usize; + let mut values = SmallVec::with_capacity(nc); - let mut values: SmallVec<[Float; 4]> = SmallVec::with_capacity(desc.offset.len()); - - match &self.pixels { - PixelData::U8(data) => { - for &channel_idx in &desc.offset { - let raw_val = data[pixel_offset + channel_idx as usize]; - let linear = self.view.encoding.to_linear_scalar(raw_val); - values.push(linear); + match &self.storage { + PixelStorage::U8(data) => { + for i in 0..nc { + values.push(self.device.base.encoding.to_linear_scalar(data[offset + i])); } } - PixelData::F16(data) => { - for &channel_idx in &desc.offset { - let raw_val = data[pixel_offset + channel_idx as usize]; - values.push(f16_to_f32(raw_val)); + PixelStorage::F16(data) => { + for i in 0..nc { + values.push(data[offset + i].to_f32()); } } - PixelData::F32(data) => { - for &channel_idx in &desc.offset { - let val = data[pixel_offset + channel_idx as usize]; - values.push(val); + PixelStorage::F32(data) => { + for i in 0..nc { + values.push(data[offset + i]); } } } @@ -419,14 +296,70 @@ impl Image { ImageChannelValues(values) } - pub fn get_channel_with_wrap( + // Write + pub fn set_channel(&mut self, p: Point2i, c: i32, mut value: Float) { + if value.is_nan() { + value = 0.0; + } + + let res = self.resolution(); + if p.x() < 0 || p.x() >= res.x() || p.y() < 0 || p.y() >= res.y() { + return; + } + + let offset = self.pixel_offset(p) + c as usize; + + match &mut self.storage { + PixelStorage::U8(data) => { + let data = Box::as_mut(data); + data[offset] = self.device.base.encoding.from_linear_scalar(value); + } + PixelStorage::F16(data) => { + let data = Box::as_mut(data); + data[offset] = f16::from_f32(value); + } + PixelStorage::F32(data) => { + let data = Box::as_mut(data); + data[offset] = value; + } + } + } + + // Descriptions + pub fn get_channels_with_desc( &self, p: Point2i, - c: i32, - wrap_mode: Option, - ) -> Float { - // if !self.remap_pixel_coords - 0. + desc: &ImageChannelDesc, + wrap_mode: WrapMode2D, + ) -> ImageChannelValues { + let mut pp = p; + if !self.device.base.remap_pixel_coords(&mut pp, wrap_mode) { + return ImageChannelValues(smallvec![0.0; desc.offset.len()]); + } + + let pixel_offset = self.pixel_offset(pp); + let mut values = SmallVec::with_capacity(desc.offset.len()); + + match &self.storage { + PixelStorage::U8(data) => { + for &c in &desc.offset { + let raw = data[pixel_offset + c]; + values.push(self.device.base.encoding.to_linear_scalar(raw)); + } + } + PixelStorage::F16(data) => { + for &c in &desc.offset { + values.push(data[pixel_offset + c].to_f32()); + } + } + PixelStorage::F32(data) => { + for &c in &desc.offset { + values.push(data[pixel_offset + c]); + } + } + } + + ImageChannelValues(values) } pub fn channel_names_from_desc(&self, desc: &ImageChannelDesc) -> Vec<&str> { @@ -449,7 +382,7 @@ impl Image { } None => { return Err(format!( - "Image is missing requested channel '{}'. Available channels: {:?}", + "Missing channel '{}'. Available: {:?}", req, self.channel_names )); } @@ -461,87 +394,53 @@ impl Image { pub fn all_channels_desc(&self) -> ImageChannelDesc { ImageChannelDesc { - offset: (0..self.n_channels()).collect(), + offset: (0..self.n_channels() as usize).collect(), } } pub fn select_channels(&self, desc: &ImageChannelDesc) -> Self { - let desc_channel_names: Vec = desc + let new_names: Vec = desc .offset .iter() - .map(|&i| self.channel_names[i as usize]) + .map(|&i| self.channel_names[i].clone()) .collect(); - let new_storage = match &self._storage { - PixelStorage::U8(src_data) => { - // Allocate destination buffer - let mut dst_data = vec![0u8; pixel_count * dst_n_channels]; + let res = self.resolution(); + let pixel_count = (res.x() * res.y()) as usize; + let src_nc = self.n_channels() as usize; + let dst_nc = desc.offset.len(); - // Iterate over every pixel (Flat loop is faster than nested x,y) + let new_storage = match &self.storage { + PixelStorage::U8(src) => { + let mut dst = vec![0u8; pixel_count * dst_nc]; for i in 0..pixel_count { - let src_pixel_start = i * src_n_channels; - let dst_pixel_start = i * dst_n_channels; - - // Copy specific channels based on desc - for (out_idx, &in_channel_offset) in desc.offset.iter().enumerate() { - let val = src_data[src_pixel_start + in_channel_offset as usize]; - dst_data[dst_pixel_start + out_idx] = val; + for (out_idx, &in_c) in desc.offset.iter().enumerate() { + dst[i * dst_nc + out_idx] = src[i * src_nc + in_c]; } } - PixelStorage::U8(dst_data) + PixelStorage::U8(dst.into_boxed_slice()) } - PixelStorage::F16(src_data) => { - let mut dst_data = vec![half::f16::ZERO; pixel_count * dst_n_channels]; - + PixelStorage::F16(src) => { + let mut dst = vec![f16::ZERO; pixel_count * dst_nc]; for i in 0..pixel_count { - let src_pixel_start = i * src_n_channels; - let dst_pixel_start = i * dst_n_channels; - - for (out_idx, &in_channel_offset) in desc.offset.iter().enumerate() { - let val = src_data[src_pixel_start + in_channel_offset as usize]; - dst_data[dst_pixel_start + out_idx] = val; + for (out_idx, &in_c) in desc.offset.iter().enumerate() { + dst[i * dst_nc + out_idx] = src[i * src_nc + in_c]; } } - PixelStorage::F16(dst_data) + PixelStorage::F16(dst.into_boxed_slice()) } - PixelStorage::F32(src_data) => { - let mut dst_data = vec![0.0; pixel_count * dst_n_channels]; - + PixelStorage::F32(src) => { + let mut dst = vec![0.0f32; pixel_count * dst_nc]; for i in 0..pixel_count { - let src_pixel_start = i * src_n_channels; - let dst_pixel_start = i * dst_n_channels; - - for (out_idx, &in_channel_offset) in desc.offset.iter().enumerate() { - let val = src_data[src_pixel_start + in_channel_offset as usize]; - dst_data[dst_pixel_start + out_idx] = val; + for (out_idx, &in_c) in desc.offset.iter().enumerate() { + dst[i * dst_nc + out_idx] = src[i * src_nc + in_c]; } } - PixelStorage::F32(dst_data) + PixelStorage::F32(dst.into_boxed_slice()) } }; - let image = Self::new( - self.format, - self.resolution, - desc_channel_names, - self.encoding(), - ); - } - - pub fn set_channels( - &mut self, - p: Point2i, - desc: &ImageChannelDesc, - mut values: &ImageChannelValues, - ) { - assert_eq!(desc.size(), values.len()); - for i in 0..desc.size() { - self.view.set_channel(p, desc.offset[i], values[i]); - } - } - - pub fn set_channels_all(&mut self, p: Point2i, values: &ImageChannelValues) { - self.set_channels(p, &self.all_channels_desc(), values) + Self::from_storage(new_storage, res, new_names, self.encoding()) } pub fn get_sampling_distribution(&self, dxd_a: F, domain: Bounds2f) -> Array2D @@ -583,9 +482,9 @@ impl Image { pub fn mse( &self, desc: ImageChannelDesc, - ref_img: &DeviceImage, + ref_img: &Image, generate_mse_image: bool, - ) -> (ImageChannelValues, Option) { + ) -> (ImageChannelValues, Option) { let mut sum_se: Vec = vec![0.; desc.size()]; let names_ref = self.channel_names_from_desc(&desc); let ref_desc = ref_img @@ -630,7 +529,7 @@ impl Image { sum_se.iter().map(|&s| (s / pixel_count) as Float).collect(); let mse_image = if generate_mse_image { - Some(DeviceImage::new( + Some(Image::new( PixelFormat::F32, self.resolution, &names_ref, @@ -668,3 +567,11 @@ impl Image { return false; } } + +impl std::ops::Deref for Image { + type Target = DeviceImage; + + fn deref(&self) -> &DeviceImage { + &self.device + } +} diff --git a/src/core/light.rs b/src/core/light.rs index 4fbf6c7..4ec1c76 100644 --- a/src/core/light.rs +++ b/src/core/light.rs @@ -1,20 +1,16 @@ -use shared::core::geometry::{Bounds3f, Point2i}; -use shared::core::image::DeviceImage; -use shared::core::light::{Light, LightBase}; -use shared::core::medium::MediumInterface; -use shared::core::spectrum::Spectrum; -use shared::core::texture::SpectrumType; -use shared::lights::{ - DistantLight, GoniometricLight, InfiniteUniformLight, PointLight, ProjectionLight, SpotLight, -}; -use shared::spectra::{DenselySampledSpectrum, RGBColorSpace, SampledSpectrum, SampledWavelengths}; -use shared::utils::{Ptr, Transform}; -use shared::{Float, PI}; - use crate::core::spectrum::{SPECTRUM_CACHE, spectrum_to_photometric}; +use crate::core::texture::FloatTexture; use crate::lights::*; use crate::utils::containers::InternCache; -use crate::utils::{Arena, ParameterDictionary, Upload, resolve_filename}; +use crate::utils::{Arena, FileLoc, ParameterDictionary, Upload, resolve_filename}; +use shared::core::camera::CameraTransform; +use shared::core::light::Light; +use shared::core::medium::Medium; +use shared::core::spectrum::Spectrum; +use shared::core::texture::SpectrumType; +use shared::lights::*; +use shared::spectra::{DenselySampledSpectrum, RGBColorSpace}; +use shared::utils::{Ptr, Transform}; pub fn lookup_spectrum(s: &Spectrum) -> DenselySampledSpectrum { let cache = SPECTRUM_CACHE.get_or_init(InternCache::new); @@ -23,21 +19,6 @@ pub fn lookup_spectrum(s: &Spectrum) -> DenselySampledSpectrum { } pub trait CreateLight { - fn new( - render_from_light: Transform, - medium_interface: MediumInterface, - le: Spectrum, - scale: Float, - shape: Option>, - alpha: Option>, - image: Option>, - image_color_space: Option>, - two_sided: Option, - fov: Option, - cos_fallof_start: Option, - total_width: Option, - ) -> Self; - fn create( arena: &mut Arena, render_from_light: Transform, @@ -45,7 +26,7 @@ pub trait CreateLight { parameters: &ParameterDictionary, loc: &FileLoc, shape: &Shape, - alpha_tex: &FloatTexture, + alpha_text: &FloatTexture, colorspace: Option<&RGBColorSpace>, ) -> Light; } @@ -61,7 +42,8 @@ pub trait LightFactory { shape: &Shape, alpha_tex: &FloatTexture, colorspace: Option<&RGBColorSpace>, - ) -> Self; + camera_transform: CameraTransform, + ) -> Result; } impl LightFactory for Light { @@ -75,8 +57,19 @@ impl LightFactory for Light { shape: &Shape, alpha_tex: &FloatTexture, colorspace: Option<&RGBColorSpace>, - ) -> Self { + camera_transform: CameraTransform, + ) -> Result { match name { + "diffuse" => lights::diffuse::create( + arena, + render_from_light, + medium, + parameters, + loc, + shape, + alpha_tex, + colorspace, + )?, "point" => PointLight::create( arena, render_from_light, @@ -86,7 +79,7 @@ impl LightFactory for Light { shape, alpha_tex, colorspace, - ), + )?, "spot" => SpotLight::create( arena, render_from_light, @@ -96,7 +89,7 @@ impl LightFactory for Light { shape, alpha_tex, colorspace, - ), + )?, "goniometric" => GoniometricLight::create( arena, render_from_light, @@ -106,7 +99,7 @@ impl LightFactory for Light { shape, alpha_tex, colorspace, - ), + )?, "projection" => ProjectionLight::create( arena, render_from_light, @@ -116,7 +109,7 @@ impl LightFactory for Light { shape, alpha_tex, colorspace, - ), + )?, "distant" => DistantLight::create( arena, render_from_light, @@ -126,66 +119,17 @@ impl LightFactory for Light { shape, alpha_tex, colorspace, - ), - "infinite" => { - let colorspace = parameters.color_space.unwrap(); - let l = parameters.get_spectrum_array("L", SpectrumType::Illuminant); - let mut scale = parameters.get_one_float("scale", 1.); - let portal = parameters.get_point3f_array("portal"); - let filename = resolve_filename(parameters.get_one_string("filename", "")); - let e_v = parameters.get_one_float("illuminance", -1.); - if l.is_empty() && filename.is_empty() && portal.is_empty() { - if e_v > 0. { - let k_e = PI; - scale *= e_v / k_e; - } - let specific = InfiniteUniformLight::new( - render_from_light, - medium_interface, - arena.alloc(le), - scale, - None, - None, - None, - None, - None, - None, - None, - None, - ); - arena.alloc(specific); - return Light::InfiniteUniform(specific); - } else if !l.is_empty() && portal.is_empty() { - if !filename.is_empty() { - panic!( - "{}: Both \"L\" and \"filename\" specified for DiffuseAreaLight.", - loc - ); - } - scale /= spectrum_to_photometric(l[0]); - if e_v > 0. { - let k_e = PI; - scale *= e_v / k_e; - } - let specific = InfiniteUniformLight::new( - render_from_light, - medium_interface, - arena.alloc(le), - scale, - None, - None, - None, - None, - None, - None, - None, - None, - ); - - arena.alloc(specific); - return Light::InfiniteUniform(specific); - } - } + )?, + "infinite" => infinite::create( + arena, + render_from_light, + medium, + camera_transform, + parameters, + colorspace, + loc, + )?, + _ => Err(error!(loc, "unknown light type: \"{}\"", name)), } } } diff --git a/src/core/material.rs b/src/core/material.rs index 873d9dd..737d2f9 100644 --- a/src/core/material.rs +++ b/src/core/material.rs @@ -8,10 +8,10 @@ use std::collections::HashMap; pub trait CreateMaterial: Sized { fn create( parameters: &TextureParameterDictionary, - normal_map: Option>, + normal_map: Option>, named_materials: &HashMap, loc: &FileLoc, - ) -> Self; + ) -> Result; } macro_rules! make_material_factory { @@ -35,10 +35,10 @@ pub trait MaterialFactory { fn create( name: &str, params: &TextureParameterDictionary, - normal_map: Arc, + normal_map: Ptr, named_materials: HashMap, loc: &FileLoc, - ) -> Result; + ) -> Result; } impl MaterialFactory for Material { diff --git a/src/core/scene.rs b/src/core/scene.rs index 80d7c6d..bb75ec5 100644 --- a/src/core/scene.rs +++ b/src/core/scene.rs @@ -6,7 +6,9 @@ use crate::utils::parameters::{ NamedTextures, ParameterDictionary, ParsedParameterVector, TextureParameterDictionary, }; use crate::utils::parser::ParserTarget; +use crate::utils::{Arena, Upload}; use crate::utils::{normalize_utf8, resolve_filename}; +use image_rs::Primitive; use parking_lot::Mutex; use shared::Float; use shared::core::camera::{Camera, CameraTransform}; @@ -16,8 +18,9 @@ use shared::core::filter::Filter; use shared::core::geometry::{Point3f, Vector3f}; use shared::core::lights::Light; use shared::core::material::Material; -use shared::core::medium::Medium; +use shared::core::medium::{Medium, MediumInterface}; use shared::core::options::RenderingCoordinateSystem; +use shared::core::primitive::{Primitive, PrimitiveTrait}; use shared::core::sampler::Sampler; use shared::core::spectrum::{Spectrum, SpectrumType}; use shared::core::texture::{FloatTexture, SpectrumTexture}; @@ -26,7 +29,6 @@ use shared::spectra::RGBColorSpace; use shared::utils::error::FileLoc; use shared::utils::math::SquareMatrix; use shared::utils::transform::{AnimatedTransform, Transform, look_at}; -use std::char::MAX; use std::collections::{HashMap, HashSet}; use std::ops::{Index as IndexTrait, IndexMut as IndexMutTrait}; use std::sync::Arc; @@ -172,6 +174,33 @@ pub struct BasicScene { pub film_state: Mutex>, } +struct SceneLookup<'a> { + textures: &'a NamedTextures, + media: &'a HashMap, + named_materials: &'a HashMap, + materials: &'a Vec, + shape_lights: &'a HashMap>, +} + +impl<'a> SceneLookup<'a> { + fn find_medium(&self, name: &str, loc: &FileLoc) -> Option { + if name.is_empty() { + return None; + } + self.media.get(name).cloned().or_else(|| { + panic!("{}: medium '{}' not defined", loc, name); + }) + } + + fn resolve_material(&self, name: &str, loc: &FileLoc) -> Material { + if !name.is_empty() { + *self.named_materials.get(name).expect("Material not found") + } else { + self.materials[index] + } + } +} + impl BasicScene { fn set_options( self: &Arc, @@ -374,7 +403,7 @@ impl BasicScene { pub fn add_light(&self, light: LightSceneEntity) { if light.transformed_base.render_from_object.is_animated() { - log::info!( + log::warn!( "{}: Animated world to texture not supported, using start", light.transformed_base.base.loc ); @@ -530,6 +559,190 @@ impl BasicScene { (named_materials_out, materials_out) } + pub fn create_aggregate( + &self, + arena: &mut Arena, + textures: &NamedTextures, + shape_lights: &HashMap>, + ) -> Arc { + let shapes_guard = self.shapes.lock().unwrap(); + let animated_shapes_guard = self.animated_shapes.lock().unwrap(); + let media_guard = self.media_state.lock().unwrap(); + let mat_guard = self.material_state.lock().unwrap(); + + let lookup = SceneLookup { + textures: &textures.float_textures, + media: &media_guard.map, + named_materials: &mat_guard.named_materials.iter().cloned().collect(), + materials: &mat_guard.materials, + shape_lights, + }; + + let mut all_primitives = Vec::new(); + + // Parallel CPU Load + let loaded_static = self.load_shapes_parallel(&shapes_guard, &lookup); + // Serial Arena Upload + let static_prims = self.upload_shapes(arena, &shapes_guard, loaded_static, &lookup); + all_primitives.extend(static_prims); + + // Parallel CPU Load + let loaded_animated = self.load_animated_shapes_parallel(&animated_shapes_guard, &lookup); + // Serial Arena Upload + let anim_prims = + self.upload_animated_shapes(arena, &animated_shapes_guard, loaded_animated, &lookup); + all_primitives.extend(anim_prims); + + // (Similar pattern: Lock definitions, load parallel, upload serial) + // Call a helper `create_instance_primitives` here. + + if !all_primitives.is_empty() { + todo!("Build BVH or KD-Tree") + } else { + todo!("Return empty") + } + } + + fn load_shapes_parallel( + &self, + entities: &[ShapeSceneEntity], + lookup: &SceneLookup, + ) -> Vec> { + entities + .par_iter() + .map(|sh| { + Shape::create( + &sh.base.name, + sh.render_from_object.as_ref(), + sh.object_from_render.as_ref(), + sh.reverse_orientation, + &sh.base.parameters, + lookup.float_textures, + &sh.base.loc, + ) + }) + .collect() + } + + fn load_animated_shapes_parallel( + &self, + entities: &[AnimatedShapeSceneEntity], + lookup: &SceneLookup, + ) -> Vec> { + entities + .par_iter() + .map(|sh| { + Shape::create( + &sh.transformed_base.base.name, + &sh.identity, + &sh.identity, + sh.reverse_orientation, + &sh.transformed_base.base.parameters, + lookup.float_textures, + &sh.transformed_base.base.loc, + ) + }) + .collect() + } + + fn upload_shapes( + &self, + arena: &mut Arena, + entities: &[ShapeSceneEntity], + loaded_shapes: Vec>, + lookup: &SceneLookup, + ) -> Vec { + let mut primitives = Vec::new(); + + for (i, (entity, shapes)) in entities.iter().zip(loaded_shapes).enumerate() { + if shapes.is_empty() { + continue; + } + + let alpha_tex = self.get_alpha_texture( + &entity.base.parameters, + &entity.base.loc, + arena, + lookup.float_textures, + ); + + let mtl = lookup.resolve_material(&entity.material, &entity.base.loc); + let mi = MediumInterface::new( + lookup.find_medium(&entity.inside_medium, &entity.base.loc), + lookup.find_medium(&entity.outside_medium, &entity.base.loc), + ); + + let shape_lights_opt = lookup.shape_lights.get(&i); + + for (j, shape_host) in shapes.into_iter().enumerate() { + let mut area_light = None; + if entity.light_index.is_some() { + if let Some(lights) = shape_lights_opt { + if j < lights.len() { + area_light = Some(lights[j]); + } + } + } + + let shape_ptr = shape_host.upload(arena); + + let prim = if area_light.is_none() + && !mi.is_medium_transition() + && alpha_tex.is_none() + { + let p = SimplePrimitive::new(shape_ptr, mtl.unwrap()); + Primitive::Simple(arena.alloc(p)) + } else { + let p = + GeometricPrimitive::new(shape_ptr, mtl.unwrap(), area_light, mi, alpha_tex); + Primitive::Geometric(arena.alloc(p)) + }; + + primitives.push(prim); + } + } + primitives + } + + fn upload_animated_shapes( + &self, + arena: &mut Arena, + entities: &[AnimatedShapeSceneEntity], + loaded_shapes: Vec>, + lookup: &SceneLookup, + ) -> Vec { + // Logic mirrors upload_shapes, but constructs AnimatedPrimitive or BVH + // ... + // Note: For AnimatedPrimitives, you wrap the result in `arena.alloc(AnimatedPrimitive::new(...))` + Vec::new() + } + + fn get_alpha_texture( + &self, + params: &ParameterDictionary, + loc: &FileLoc, + arena: &mut Arena, + textures: &HashMap, + ) -> Option { + let alpha_name = params.get_texture("alpha"); + + if let Some(name) = alpha_name { + match textures.get(&name) { + Some(tex) => Some(*tex), + None => panic!("{:?}: Alpha texture '{}' not found", loc, name), + } + } else { + let alpha_val = params.get_one_float("alpha", 1.0); + if alpha_val < 1.0 { + let tex = FloatConstantTexture::new(alpha_val); + let ptr = arena.alloc(tex); + Some(FloatTexture::Constant(ptr)) + } else { + None + } + } + } + pub fn get_camera(&self) -> Arc { self.get_singleton(&self.camera_state, "Camera") } @@ -650,7 +863,6 @@ impl BasicScene { } // PRIVATE METHODS - fn get_singleton( &self, mutex: &Mutex>, @@ -852,7 +1064,6 @@ impl BasicSceneBuilder { named_coordinate_systems: HashMap::new(), active_instance_definition: None, shapes: Vec::new(), - instances: Vec::new(), named_material_names: HashSet::new(), medium_names: HashSet::new(), diff --git a/src/lights/diffuse.rs b/src/lights/diffuse.rs index 4d197fa..a8ce937 100644 --- a/src/lights/diffuse.rs +++ b/src/lights/diffuse.rs @@ -6,6 +6,7 @@ use shared::core::spectrum::{Spectrum, SpectrumTrait}; use shared::core::texture::SpectrumType; use shared::lights::DiffuseAreaLight; use shared::spectra::RGBColorSpace; +use shared::utils::Transform; use std::sync::Arc; use crate::core::image::{Image, ImageIO}; @@ -17,18 +18,33 @@ use crate::core::texture::{ }; use crate::utils::{Arena, FileLoc, ParameterDictionary, Upload, resolve_filename}; -impl CreateLight for DiffuseAreaLight { +pub trait CreateDiffuseLight { fn new( render_from_light: shared::utils::Transform, medium_interface: MediumInterface, le: Spectrum, scale: Float, - shape: Option>, - alpha: Option>, - image: Option>, - image_color_space: Option>, - two_sided: Option, - fov: Option, + shape: Ptr, + alpha: Ptr, + image: Ptr, + colorspace: Option, + two_sided: bool, + fov: Float, + ) -> Self; +} + +impl CreateDiffuseLight for DiffuseAreaLight { + fn new( + render_from_light: shared::utils::Transform, + medium_interface: MediumInterface, + le: Spectrum, + scale: Float, + shape: Ptr, + alpha: Ptr, + image: Ptr, + colorspace: Option, + two_sided: bool, + fov: Float, ) -> Self { let is_constant_zero = match &alpha { FloatTexture::FloatConstant(tex) => tex.evaluate(&TextureEvalContext::default()) == 0.0, @@ -56,7 +72,7 @@ impl CreateLight for DiffuseAreaLight { ); assert!( - image_color_space.is_some(), + colorspace.is_some(), "Image provided but ColorSpace is missing" ); } @@ -74,7 +90,7 @@ impl CreateLight for DiffuseAreaLight { base, area: shape.area(), image, - image_color_space, + colorspace, shape: Ptr::from(&*storage.shape), alpha: stored_alpha, lemit, @@ -82,7 +98,9 @@ impl CreateLight for DiffuseAreaLight { scale, } } +} +impl CreateLight for DiffuseAreaLight { fn create( arena: &mut Arena, render_from_light: Transform, @@ -92,40 +110,44 @@ impl CreateLight for DiffuseAreaLight { shape: &Shape, alpha_text: &FloatTexture, colorspace: Option<&RGBColorSpace>, - ) -> Light { + ) -> Result { let mut l = params.get_one_spectrum("l", None, SpectrumType::Illuminant); let mut scale = params.get_one_float("scale", 1.); let two_sided = params.get_one_bool("twosided", false); + let filename = resolve_filename(params.get_one_string("filename", "")); - let mut image_color_space = colorspace.clone(); - if !filename.is_empty() { + let (image, image_color_space) = if !filename.is_empty() { if l.is_some() { - panic!( - "{}: Both \"L\" and \"filename\" specified for DiffuseAreaLight.", - loc - ); + return Err(error!(loc, "both \"L\" and \"filename\" specified")); } - let im = Image::read(filename, None); - let image: Image = im.image; - if image.has_any_infinite_pixels() { - panic!( - "{:?}: Image '{}' has NaN pixels. Not suitable for light.", - loc, filename - ); + + let im = Image::read(&filename, None)?; + + if im.image.has_any_infinite_pixels() { + return Err(error!(loc, "{}: image has infinite pixel values", filename)); } - let channel_desc = image.get_channel_desc(&["R", "G", "B"]).expect(&format!( - "{:?}: Image '{}' must have R, G, B channels", - loc, filename - )); - let selected_image = image.select_channels(channel_desc); - image_color_space = im.metadata.colorspace; - } else if l.is_none() { - l = Some(colorpace.unwrap().Illuminant.clone()); - } + if im.image.has_any_nan_pixels() { + return Err(error!(loc, "{}: image has NaN pixel values", filename)); + } + + let channel_desc = im + .image + .get_channel_desc(&["R", "G", "B"]) + .ok_or_else(|| error!(loc, "{}: image must have R, G, B channels", filename))?; + + let image = im.image.select_channels(&channel_desc); + let cs = im.metadata.get_color_space(); + + (Some(image), Some(cs)) + } else { + if l.is_none() { + l = Some(colorspace.illuminant.clone()); + } + (None, None) + }; let l_for_scale = l.as_ref().unwrap_or(&colorspace.illuminant); - let lemit_data = lookup_spectrum(l_for_scale); - scale /= spectrum_to_photometric(lemit_data); + scale /= spectrum_to_photometric(l_for_scale); let phi_v = parameters.get_one_float("power", -1.0); if phi_v > 0.0 { @@ -165,16 +187,16 @@ impl CreateLight for DiffuseAreaLight { let specific = DiffuseAreaLight::new( render_from_light, medium.into(), - l, + l.as_ref(), scale, shape.upload(arena), alpha.upload(arena), image.upload(arena), image_color_space.upload(arena), - Some(true), + true, shape.area(), ); - Light::DiffuseArea(specific) + Ok(Light::DiffuseArea(specific)) } } diff --git a/src/lights/distant.rs b/src/lights/distant.rs index 8e0ded7..1a23469 100644 --- a/src/lights/distant.rs +++ b/src/lights/distant.rs @@ -8,21 +8,14 @@ use shared::core::medium::{Medium, MediumInterface}; use shared::core::shape::Shape; use shared::lights::DistantLight; use shared::spectra::RGBColorSpace; -use shared::utils::Transform; +use shared::utils::{Ptr, Transform}; -impl CreateLight for DistantLight { - fn new( - render_from_light: Transform, - medium_interface: shared::core::medium::MediumInterface, - le: shared::core::spectrum::Spectrum, - scale: shared::Float, - shape: Option>, - alpha: Option>, - image: Option>, - image_color_space: Option>, - two_sided: Option, - fov: Option, - ) -> Self { +pub trait CreateDistantLight { + fn new(render_from_light: Transform, le: Spectrum, scale: Float) -> Self; +} + +impl CreateDistantLight for DistantLight { + fn new(render_from_light: Transform, le: Spectrum, scale: Float) -> Self { let base = LightBase::new( LightType::DeltaDirection, render_from_light, @@ -37,7 +30,9 @@ impl CreateLight for DistantLight { scene_radius: 0., } } +} +impl CreateLight for DistantLight { fn create( arena: &mut Arena, render_from_light: Transform, @@ -47,7 +42,7 @@ impl CreateLight for DistantLight { shape: &Shape, alpha_text: &FloatTexture, colorspace: Option<&RGBColorSpace>, - ) -> Light { + ) -> Result { let l = parameters .get_one_spectrum( "L", @@ -55,7 +50,6 @@ impl CreateLight for DistantLight { SpectrumType::Illuminant, ) .unwrap(); - let lemit = lookup_spectrum(l); let mut scale = parameters.get_one_float("scale", 1); let from = parameters.get_one_point3f("from", Point3f::new(0., 0., 0.)); @@ -90,19 +84,9 @@ impl CreateLight for DistantLight { if e_v > 0. { sc *= e_v; } - let specific = DistantLight::new( - final_render, - medium.into(), - le, - scale, - None, - None, - None, - None, - None, - None, - ); - Light::Distant(specific) + let specific = DistantLight::new(final_render, l, scale); + + Ok(Light::Distant(specific)) } } diff --git a/src/lights/goniometric.rs b/src/lights/goniometric.rs index e8107f1..c0bb752 100644 --- a/src/lights/goniometric.rs +++ b/src/lights/goniometric.rs @@ -2,8 +2,10 @@ use crate::core::image::{Image, ImageIO, ImageMetadata}; use crate::core::light::{CreateLight, lookup_spectrum}; use crate::core::spectrum::spectrum_to_photometric; use crate::core::texture::FloatTexture; +use crate::lights::distant::CreateDistantLight; use crate::utils::sampling::PiecewiseConstant2D; use crate::utils::{Arena, FileLoc, ParameterDictionary, resolve_filename}; +use shared::Float; use shared::core::image::ImageBase; use shared::core::light::{Light, LightBase, LightType}; use shared::core::medium::{Medium, MediumInterface}; @@ -11,21 +13,26 @@ use shared::core::spectrum::Spectrum; use shared::core::texture::SpectrumType; use shared::lights::GoniometricLight; use shared::spectra::RGBColorSpace; -use shared::utils::Transform; use shared::utils::containers::Array2D; +use shared::utils::{Ptr, Transform}; -impl CreateLight for GoniometricLight { +pub trait CreateGoniometricLight { fn new( render_from_light: Transform, medium_interface: MediumInterface, le: Spectrum, - scale: shared::Float, - shape: Option>, - alpha: Option>, - image: Option>, - image_color_space: Option>, - two_sided: Option, - fov: Option, + scale: Float, + image: Ptr, + ) -> Self; +} + +impl CreateGoniometricLight for GoniometricLight { + fn new( + render_from_light: Transform, + medium_interface: MediumInterface, + le: Spectrum, + scale: Float, + image: Ptr, ) -> Self { let base = LightBase::new( LightType::DeltaPosition, @@ -44,7 +51,9 @@ impl CreateLight for GoniometricLight { distrib, } } +} +impl CreateLight for GoniometricLight { fn create( arena: &mut Arena, render_from_light: Transform, @@ -54,96 +63,50 @@ impl CreateLight for GoniometricLight { shape: &Shape, alpha_text: &FloatTexture, colorspace: Option<&RGBColorSpace>, - ) -> Light { + ) -> Result { let i = params.get_one_spectrum( "I", colorspace.unwrap().illuminant, SpectrumType::Illuminant, ); - let lemit_data = lookup_spectrum(&i_spectrum_def); let sc = params.get_one_float("scale", 1.); let filename = resolve_filename(params.get_one_string("filename", "")); let mut image: Option = None; - - if filename.is_empty() { - println!( - "{}: Both \"L\" and \"filename\" specified for DiffuseAreaLight.", - loc - ) + let image = if filename.is_empty() { + None } else { - let im = Image::read(filename, None).expect("Could not load image"); - let loaded_img: Image = im.image; - let res = loaded_img.resolution(); - let metadata: ImageMetadata = im.metadata; - if loaded_img.has_any_infinite_pixels() { - panic!( - "{:?}: Image '{}' has NaN pixels. Not suitable for light.", - loc, filename - ); - } - if res.x() != res.y() { - panic!( - "{:?}: Image resolution ({}, {}) is non square. Unlikely that this is an equal area map.", + let im = Image::read(&filename, None) + .map_err(|e| error!(loc, "could not load image '{}': {}", filename, e))?; + + let loaded = im.image; + let res = loaded.resolution(); + + if loaded.has_any_infinite_pixels() { + return Err(error!( loc, - res.x(), - res.y(), - y() - ); - } - let rgb_desc = loaded_img.get_channel_desc(&["R", "G", "B"]); - let y_desc = loaded_img.get_channel_desc(&["Y"]); - if let Ok(rgb) = rgb_desc { - if y_desc.is_ok() { - panic!( - "{:?}: Image '{}' has both RGB and Y channels. Ambiguous.", - loc, filename - ); - } - - let mut y_pixels = Vec::with_capacity((res.x * res.y) as usize); - - for y in 0..res.y { - for x in 0..res.x { - let r = loaded_img.get_channel(Point2i::new(x, y), 0); - let g = loaded_img.get_channel(Point2i::new(x, y), 1); - let b = loaded_img.get_channel(Point2i::new(x, y), 2); - - y_pixels.push((r + g + b) / 3.0); - } - } - - loaded_img = Some(Image::new( - PixelFormat::F32, - res, - &["Y"], - ColorEncoding::Linear, + "image '{}' has infinite pixels, not suitable for light", filename )); - } else if y_desc.is_ok() { - image = Some(loaded_img); - } else { - panic!( - "{:?}: Image '{}' has neither RGB nor Y channels.", - loc, filename - ); } - } + + if res.x != res.y { + return Err(error!( + loc, + "image resolution ({}, {}) is non-square; unlikely to be an equal-area map", + res.x, + res.y + )); + } + + Some(convert_to_luminance_image(&loaded, &filename, loc)?) + }; scale /= spectrum_to_photometric(&lemit_data); let phi_v = params.get_one_float("power", -1.0); if phi_v > 0.0 { if let Some(ref img) = image { - let mut sum_y = 0.0; - let res = img.resolution(); - - for y in 0..res.y { - for x in 0..res.x { - sum_y += img.get_channel(Point2i::new(x, y), 0); - } - } - - let k_e = 4.0 * PI * sum_y / (res.x * res.y) as Float; - scale *= phi_v / k_e; + let k_e = compute_emissive_power(image); + scale *= phi_v / phi_e; } } @@ -153,22 +116,71 @@ impl CreateLight for GoniometricLight { let t = Transform::from_flat(swap_yz); let final_render_from_light = render_from_light * t; - let d: Array2D = image.get_sampling_distribution(); - distrib = PiecewiseConstant2D::new_with_data(d); + let specific = + GoniometricLight::new(final_render_from_light, medium.into(), le, scale, image); - let specific = GoniometricLight::new( - render_from_light, - medium_interface, - le, - scale, - None, - None, - Some(image), - None, - None, - None, - ); - - Light::Goniometric(specific) + Ok(Light::Goniometric(specific)) } } + +fn convert_to_luminance_image( + image: &Image, + filename: &str, + loc: &FileLoc, +) -> Result { + let res = image.resolution(); + let rgb_desc = image.get_channel_desc(&["R", "G", "B"]); + let y_desc = image.get_channel_desc(&["Y"]); + + match (rgb_desc, y_desc) { + (Ok(_), Ok(_)) => Err(error!( + loc, + "image '{}' has both RGB and Y channels; ambiguous", filename + )), + + (Ok(_), Err(_)) => { + // Convert RGB to Y (luminance) + let mut y_pixels = Vec::with_capacity((res.x * res.y) as usize); + + for y in 0..res.y { + for x in 0..res.x { + let r = image.get_channel(Point2i::new(x, y), 0); + let g = image.get_channel(Point2i::new(x, y), 1); + let b = image.get_channel(Point2i::new(x, y), 2); + y_pixels.push((r + g + b) / 3.0); + } + } + + Ok(Image::from_pixels( + PixelFormat::F32, + res, + &["Y"], + ColorEncoding::Linear, + &y_pixels, + )) + } + + (Err(_), Ok(_)) => { + // Already has Y channel, use as-is + Ok(image.clone()) + } + + (Err(_), Err(_)) => Err(error!( + loc, + "image '{}' has neither RGB nor Y channels", filename + )), + } +} + +fn compute_emissive_power(image: &Image) -> Float { + let res = image.resolution(); + let mut sum_y = 0.0; + + for y in 0..res.y { + for x in 0..res.x { + sum_y += image.get_channel(Point2i::new(x, y), 0); + } + } + + 4.0 * PI * sum_y / (res.x * res.y) as Float +} diff --git a/src/lights/infinite.rs b/src/lights/infinite.rs index e95b00d..1bcec19 100644 --- a/src/lights/infinite.rs +++ b/src/lights/infinite.rs @@ -2,26 +2,32 @@ use shared::Float; use shared::core::geometry::{Bounds3f, Point2f}; use shared::core::light::{CreateLight, Light, LightBase, LightType}; use shared::core::medium::MediumInterface; -use shared::lights::{InfiniteImageLight, InfinitePortalLight, InfiniteUniformLight}; +use shared::core::spectrum::Spectrum; +use shared::lights::{ImageInfiniteLight, PortalInfiniteLight, UniformInfiniteLight}; use shared::spectra::RGBColorSpace; -use shared::utils::Transform; use shared::utils::sampling::DevicePiecewiseConstant2D; +use shared::utils::{Ptr, Transform}; use std::sync::Arc; use crate::core::light::{LightBaseTrait, lookup_spectrum}; -impl CreateLight for InfiniteImageLight { +pub trait CreateImageInfiniteLight { fn new( render_from_light: Transform, medium_interface: MediumInterface, - le: shared::core::spectrum::Spectrum, scale: Float, - shape: Option>, - alpha: Option>, - image: Option>, - image_color_space: Option>, - two_sided: Option, - fov: Option, + image: Ptr, + image_color_space: Ptr, + ) -> Self; +} + +impl CreateImageInfiniteLight for ImageInfiniteLight { + fn new( + render_from_light: Transform, + medium_interface: MediumInterface, + scale: Float, + image: Ptr, + image_color_space: Ptr, ) -> Self { let base = LightBase::new( LightType::Infinite, @@ -65,7 +71,7 @@ impl CreateLight for InfiniteImageLight { let compensated_distrib = DevicePiecewiseConstant2D::new_with_bounds(&d, domain); - InfiniteImageLight { + ImageInfiniteLight { base, image: &image, image_color_space: &storage.image_color_space, @@ -76,19 +82,6 @@ impl CreateLight for InfiniteImageLight { compensated_distrib: &compensated_distrib, }; } - - fn create( - arena: &mut crate::utils::Arena, - render_from_light: Transform, - medium: Medium, - parameters: &crate::utils::ParameterDictionary, - loc: &FileLoc, - shape: &Shape, - alpha_tex: &FloatTexture, - colorspace: Option<&RGBColorSpace>, - ) -> shared::core::light::Light { - todo!() - } } #[derive(Debug)] @@ -99,24 +92,29 @@ struct InfinitePortalLightStorage { } #[derive(Clone, Debug)] -pub struct InfinitePortalLightHost { - pub view: InfinitePortalLight, +pub struct PortalInfiniteLightHost { + pub view: PortalInfiniteLight, pub filename: String, _storage: Arc, } -impl CreateLight for InfinitePortalLightHost { +pub trait CreatePortalInfiniteLight { fn new( render_from_light: Transform, - medium_interface: MediumInterface, - le: shared::core::spectrum::Spectrum, scale: Float, - shape: Option>, - alpha: Option>, - image: Option>, - image_color_space: Option>, - two_sided: Option, - fov: Option, + image: Ptr, + image_color_space: Ptr, + points: Ptr, + ) -> Self; +} + +impl CreatePortalInfiniteLight for PortalInfiniteLight { + fn new( + render_from_light: Transform, + scale: Float, + image: Ptr, + image_color_space: Ptr, + points: Ptr, ) -> Self { let base = LightBase::new( LightType::Infinite, @@ -124,7 +122,7 @@ impl CreateLight for InfinitePortalLightHost { &MediumInterface::default(), ); - let desc = equal_area_image + let desc = image .get_channel_desc(&["R", "G", "B"]) .unwrap_or_else(|_| { panic!( @@ -225,16 +223,10 @@ impl CreateLight for InfinitePortalLightHost { let distribution = WindowedPiecewiseConstant2D::new(d); - let storage = Arc::new(InfinitePortalLightStorage { - image, - distribution, - image_color_space, - }); - - InfinitePortalLight { + PortalInfiniteLight { base, image, - image_color_space: &storage.image_color_space, + image_color_space: &image_color_space, scale, scene_center: Point3f::default(), scene_radius: 0., @@ -243,36 +235,14 @@ impl CreateLight for InfinitePortalLightHost { distribution, } } - - fn create( - arena: &mut crate::utils::Arena, - render_from_light: Transform, - medium: Medium, - parameters: &crate::utils::ParameterDictionary, - loc: &FileLoc, - shape: &Shape, - alpha_tex: &FloatTexture, - colorspace: Option<&RGBColorSpace>, - ) -> Light { - todo!() - } } -impl CreateLight for InfiniteUniformLight { - fn new( - render_from_light: Transform, - medium_interface: MediumInterface, - le: shared::core::spectrum::Spectrum, - scale: Float, - shape: Option>, - alpha: Option>, - image: Option>, - image_color_space: Option>, - two_sided: Option, - fov: Option, - cos_fallof_start: Option, - total_width: Option, - ) -> Self { +pub trait CreateUniformInfiniteLight { + fn new(render_from_light: Transform, le: Spectrum, scale: Float) -> Self; +} + +impl CreateUniformInfiniteLight for UniformInfiniteLight { + fn new(render_from_light: Transform, le: Spectrum, scale: Float) -> Self { let base = LightBase::new( LightType::Infinite, &render_from_light, @@ -287,17 +257,135 @@ impl CreateLight for InfiniteUniformLight { scene_radius: 0., } } +} - fn create( - arena: &mut crate::utils::Arena, - render_from_light: Transform, - medium: Medium, - parameters: &crate::utils::ParameterDictionary, - loc: &FileLoc, - shape: &Shape, - alpha_tex: &FloatTexture, - colorspace: Option<&RGBColorSpace>, - ) -> Light { - todo!() +pub fn create( + arena: &mut Arena, + render_from_light: Transform, + medium: MediumInterface, + camera_transform: CameraTransform, + parameters: &ParameterDictionary, + colorspace: &RGBColorSpace, + loc: &FileLoc, +) -> Result { + let l = parameters.get_spectrum_array("L", SpectrumType::Illuminant); + let mut scale = parameters.get_one_float("scale", 1.0); + let portal = parameters.get_point3f_array("portal"); + let filename = resolve_filename(parameters.get_one_string("filename", "")); + let e_v = parameters.get_one_float("illuminance", -1.0); + + let has_spectrum = !l.is_empty(); + let has_file = !filename.is_empty(); + let has_portal = !portal.is_empty(); + + if has_spectrum && has_file { + return Err(error!(loc, "cannot specify both \"L\" and \"filename\"")); + } + + if !has_file && !has_portal { + let spectrum = if has_spectrum { + scale /= spectrum_to_photometric(&l[0]); + &l[0] + } else { + &colorspace.illuminant + }; + + if e_v > 0.0 { + scale *= e_v / PI; + } + + let light = UniformInfiniteLight::new(render_from_light, lookup_spectrum(spectrum), scale); + return Ok(Light::InfiniteUniform(light)); + } + + // Image based + + let (image, image_cs) = load_image_or_constant(&filename, &l, colorspace, loc)?; + + scale /= spectrum_to_photometric(&image_cs.illuminant); + + if e_v > 0.0 { + let k_e = compute_hemisphere_illuminance(&image, &image_cs); + scale *= e_v / k_e; + } + + let image_ptr = image.upload(arena); + let cs_ptr = image_cs.upload(arena); + + if has_portal { + let portal_render: Vec = portal + .iter() + .map(|p| camera_transform.render_from_world(*p)) + .collect(); + + let light = PortalInfiniteLight::new( + render_from_light, + scale, + image_ptr, + cs_ptr, + arena.alloc_slice(&portal_render), + ); + Ok(Light::InfinitePortal(light)) + } else { + let light = ImageInfiniteLight::new(render_from_light, medium, scale, image_ptr, cs_ptr); + Ok(Light::InfiniteImage(light)) } } + +fn load_image_or_constant( + filename: &str, + l: &[Spectrum], + colorspace: &RGBColorSpace, + loc: &FileLoc, +) -> Result<(Image, RGBColorSpace), Error> { + if filename.is_empty() { + let rgb = spectrum_to_rgb(&l[0], colorspace); + let image = Image::new_constant(Point2i::new(1, 1), &["R", "G", "B"], &rgb); + Ok((image, colorspace.clone())) + } else { + let im = Image::read(filename, None) + .map_err(|e| error!(loc, "failed to load '{}': {}", filename, e))?; + + if im.image.has_any_infinite_pixels() || im.image.has_any_nan_pixels() { + return Err(error!(loc, "image '{}' has invalid pixels", filename)); + } + + im.image + .get_channel_desc(&["R", "G", "B"]) + .map_err(|_| error!(loc, "image '{}' must have R, G, B channels", filename))?; + + let cs = im + .metadata + .color_space + .unwrap_or_else(|| colorspace.clone()); + let selected = im.image.select_channels(&["R", "G", "B"])?; + + Ok((selected, cs)) + } +} + +fn compute_hemisphere_illuminance(image: &Image, cs: &RGBColorSpace) -> Float { + let lum = cs.luminance_vector(); + let res = image.resolution(); + let mut sum = 0.0; + + for y in 0..res.y { + let v = (y as Float + 0.5) / res.y as Float; + for x in 0..res.x { + let u = (x as Float + 0.5) / res.x as Float; + let w = equal_area_square_to_sphere(Point2f::new(u, v)); + + if w.z <= 0.0 { + continue; + } + + let r = image.get_channel(Point2i::new(x, y), 0); + let g = image.get_channel(Point2i::new(x, y), 1); + let b = image.get_channel(Point2i::new(x, y), 2); + + sum += (r * lum[0] + g * lum[1] + b * lum[2]) * cos_theta(&w); + } + } + + sum * 2.0 * PI / (res.x * res.y) as Float +} diff --git a/src/lights/mod.rs b/src/lights/mod.rs index cbcaadc..8445aa2 100644 --- a/src/lights/mod.rs +++ b/src/lights/mod.rs @@ -6,3 +6,13 @@ pub mod point; pub mod projection; pub mod sampler; pub mod spot; + +pub use diffuse::CreateDiffuseLight; +pub use distant::CreateDistantLight; +pub use goniometric::CreateGoniometricLight; +pub use infinite::{ + CreateImageInfiniteLight, CreatePortalInfiniteLight, CreateUniformInfiniteLight, +}; +pub use point::CreatePointLight; +pub use projection::CreateProjectionLight; +pub use spot::CreateSpotLight; diff --git a/src/lights/point.rs b/src/lights/point.rs index 18a6779..9aaf4c5 100644 --- a/src/lights/point.rs +++ b/src/lights/point.rs @@ -2,27 +2,31 @@ use crate::core::light::{CreateLight, LightBaseTrait, lookup_spectrum}; use crate::core::spectrum::spectrum_to_photometric; use crate::core::texture::FloatTexture; use crate::utils::{Arena, FileLoc, ParameterDictionary}; -use shared::PI; use shared::core::geometry::VectorLike; use shared::core::light::{Light, LightBase, LightType}; -use shared::core::medium::Medium; +use shared::core::medium::{Medium, MediumInterface}; use shared::core::shape::Shape; +use shared::core::spectrum::Spectrum; use shared::lights::PointLight; use shared::spectra::RGBColorSpace; use shared::utils::Transform; +use shared::{Float, PI}; -impl CreateLight for PointLight { +pub trait CreatePointLight { fn new( render_from_light: Transform, - medium_interface: shared::core::medium::MediumInterface, - le: shared::core::spectrum::Spectrum, - scale: shared::Float, - _shape: Option>, - _alpha: Option>, - _image: Option>, - _image_color_space: Option>, - _two_sided: Option, - _fov: Option, + medium_interface: MediumInterface, + le: Spectrum, + scale: Float, + ) -> Self; +} + +impl CreatePointLight for PointLight { + fn new( + render_from_light: Transform, + medium_interface: MediumInterface, + le: Spectrum, + scale: Float, ) -> Self { let base = LightBase::new( LightType::DeltaPosition, @@ -33,7 +37,9 @@ impl CreateLight for PointLight { Self { base, scale, i } } +} +impl CreateLight for PointLight { fn create( arena: &mut Arena, render_from_light: Transform, @@ -43,7 +49,7 @@ impl CreateLight for PointLight { shape: &Shape, alpha_text: &FloatTexture, colorspace: Option<&RGBColorSpace>, - ) -> Light { + ) -> Result { let l = parameters .get_one_spectrum( "L", @@ -51,7 +57,6 @@ impl CreateLight for PointLight { SpectrumType::Illuminant, ) .unwrap(); - let lemit = lookup_spectrum(l); let mut scale = parameters.get_one_float("scale", 1); scale /= spectrum_to_photometric(l.unwrap()); let phi_v = parameters.get_one_float("power", 1.); @@ -63,12 +68,7 @@ impl CreateLight for PointLight { let from = parameters.get_one_point3f("from", Point3f::zero()); let tf = Transform::translate(from.into()); let final_render = render_from_light * tf; - let base = LightBase::new(LightType::DeltaPosition, render_from_light, medium.into()); - let specific = PointLight { - base, - scale, - i: arena.alloc(lemit), - }; - Light::Point(specific) + let specific = PointLight::new(final_render, medium.into(), l, scale); + Ok(Light::Point(specific)) } } diff --git a/src/lights/projection.rs b/src/lights/projection.rs index 3a487fc..0b4569b 100644 --- a/src/lights/projection.rs +++ b/src/lights/projection.rs @@ -2,7 +2,8 @@ use crate::core::image::{Image, ImageIO, ImageMetadata}; use crate::core::light::CreateLight; use crate::core::spectrum::spectrum_to_photometric; use crate::spectra::colorspace::new; -use crate::utils::{Ptr, Upload, resolve_filename}; +use crate::utils::{Arena, ParameterDictionary, Ptr, Upload, resolve_filename}; +use shared::Float; use shared::core::geometry::{Bounds2f, VectorLike}; use shared::core::image::ImageAccess; use shared::core::light::{Light, LightBase}; @@ -13,18 +14,27 @@ use shared::spectra::RGBColorSpace; use shared::utils::math::{radians, square}; use shared::utils::{Ptr, Transform}; -impl CreateLight for ProjectionLight { +pub trait CreateProjectionLight { fn new( render_from_light: Transform, medium_interface: MediumInterface, le: Spectrum, - scale: shared::Float, - shape: Option>, - alpha: Option>, - image: Option>, - image_color_space: Option>, - two_sided: Option, - fov: Option, + scale: Float, + image: Ptr, + image_color_space: Ptr, + fov: Float, + ) -> Self; +} + +impl CreateProjectionLight for ProjectionLight { + fn new( + render_from_light: Transform, + medium_interface: MediumInterface, + le: Spectrum, + scale: Float, + image: Ptr, + image_color_space: Ptr, + fov: Float, ) -> Self { let base = LightBase::new( LightType::DeltaPosition, @@ -69,96 +79,122 @@ impl CreateLight for ProjectionLight { a, } } +} +impl CreateLight for ProjectionLight { fn create( - arena: &mut crate::utils::Arena, - render_from_light: shared::utils::Transform, + arena: &mut Arena, + render_from_light: Transform, medium: Medium, - parameters: &crate::utils::ParameterDictionary, + parameters: &ParameterDictionary, loc: &FileLoc, _shape: &Shape, _alpha_text: &FloatTexture, - _colorspace: Option<&shared::spectra::RGBColorSpace>, - ) -> Light { + colorspace: Option<&RGBColorSpace>, + ) -> Result { let mut scale = parameters.get_one_float("scale", 1.); let power = parameters.get_one_float("power", -1.); let fov = parameters.get_one_float("fov", 90.); + let filename = resolve_filename(parameters.get_one_string("filename", "")); if filename.is_empty() { - panic!( - "{}: Must provide filename for projection light source.", - loc - ); + return Err(error!(loc, "must provide filename for projection light")); } - let im = Image::read(filename, None).unwrap(); - let image: Image = im.image; - let metadata: ImageMetadata = im.metadata; - if image.has_any_infinite_pixels() { - panic!( - "{:?}: Image '{}' has NaN pixels. Not suitable for light.", - loc, filename - ); + + let im = Image::read(&filename, None) + .map_err(|e| error!(loc, "could not load image '{}': {}", filename, e))?; + + if im.image.has_any_infinite_pixels() { + return Err(error!( + loc, + "image '{}' has infinite pixels, not suitable for light", filename + )); } - let colorspace: RGBColorSpace = metadata.colorspace.unwrap(); - let channel_desc = image + + if im.image.has_any_nan_pixels() { + return Err(error!( + loc, + "image '{}' has NaN pixels, not suitable for light", filename + )); + } + + let channel_desc = im + .image .get_channel_desc(&["R", "G", "B"]) - .unwrap_or_else(|_| { - panic!( - "{:?}: Image '{}' must have R, G and B channels.", - loc, filename - ) - }); - let image = image.select_channels(channel_desc); + .map_err(|_| error!(loc, "image '{}' must have R, G, B channels", filename))?; + + let image = im.image.select_channels(&channel_desc); + let colorspace = im + .metadata + .colorspace + .ok_or_else(|| error!(loc, "image '{}' missing colorspace metadata", filename))?; + scale /= spectrum_to_photometric(colorspace.illuminant); if power > 0. { - let hither = 1e-3; - let aspect = image.resolution().x() as Float / image.resolution().y() as Float; - let screen_bounds = if aspect > 1. { - Bounds2f::from_poins(Point2f::new(-aspect, -1.), Point2f::new(aspect, 1.)) - } else { - Bounds2f::from_points( - Point2f::new(-1., -1. / aspect), - Point2f::new(1., 1. / aspect), - ) - }; - let screen_from_light = Transform::perspective(fov, hither, 1e30); - let light_from_screen = screen_from_light.inverse(); - let opposite = (radians(fov) / 2.).tan(); - let aspect_factor = if aspect > 1. { aspect } else { 1. / aspect }; - let a = 4. * square(aspect_factor); - let mut sum = 0; - let luminance = colorspace.luminance_vector(); - let res = image.resolution(); - for y in 0..res.y() { - for x in 0..res.x() { - let lerp_factor = Point2f::new((x + 0.5) / res.x(), (y + 0.5) / res.y()); - let ps = screen_bounds.lerp(lerp_factor); - let w_point = - light_from_screen.apply_to_point(Point3f::new(ps.x(), ps.y(), 0.)); - let w = Vector3f::from(w_point).normalize(); - let dwda = w.z().powi(3); - for c in 0..3 { - sum += image.get_channel(Point2f::new(x, y), c) * luminance[c] * dwda; - } - } - } - scale *= power / (a * sum / (res.x() + res.y())); + let k_e = compute_emissive_power(&image, colorspace, fov); } let flip = Transform::scale(1., -1., 1.); let render_from_light_flip = render_from_light * flip; - let specific = Self::new( - render_from_light, + + let specific = ProjectionLight::new( + render_from_light_flip, medium_interface, le, scale, - None, - None, image.upload(arena), - None, - None, - None, + colorspace.upload(arena), + fov, ); - Light::Projection(specific) + + Ok(Light::Projection(specific)) } } + +fn compute_screen_bounds(aspect: Float) -> Bounds2f { + if aspect > 1.0 { + Bounds2f::from_points(Point2f::new(-aspect, -1.0), Point2f::new(aspect, 1.0)) + } else { + Bounds2f::from_points( + Point2f::new(-1.0, -1.0 / aspect), + Point2f::new(1.0, 1.0 / aspect), + ) + } +} + +fn compute_emissive_power(image: &Image, colorspace: &RGBColorSpace, fov: Float) -> Float { + let res = image.resolution(); + let aspect = res.x() as Float / res.y() as Float; + let screen_bounds = compute_screen_bounds(aspect); + + let hither = 1e-3; + let screen_from_light = + Transform::perspective(fov, hither, 1e30).expect("Failed to create perspective transform"); + let light_from_screen = screen_from_light.inverse(); + + let opposite = (radians(fov) / 2.0).tan(); + let aspect_factor = if aspect > 1.0 { aspect } else { 1.0 / aspect }; + let a = 4.0 * square(opposite) * aspect_factor; + + let luminance = colorspace.luminance_vector(); + let mut sum: Float = 0.0; + + for y in 0..res.y() { + for x in 0..res.x() { + let lerp_factor = Point2f::new( + (x as Float + 0.5) / res.x() as Float, + (y as Float + 0.5) / res.y() as Float, + ); + let ps = screen_bounds.lerp(lerp_factor); + let w_point = light_from_screen.apply_to_point(Point3f::new(ps.x(), ps.y(), 0.0)); + let w = Vector3f::from(w_point).normalize(); + let dwda = w.z().powi(3); + + for c in 0..3 { + sum += image.get_channel(Point2i::new(x, y), c) * luminance[c] * dwda; + } + } + } + + a * sum / (res.x() * res.y()) as Float +} diff --git a/src/lights/spot.rs b/src/lights/spot.rs index b694922..f944154 100644 --- a/src/lights/spot.rs +++ b/src/lights/spot.rs @@ -2,8 +2,7 @@ use crate::core::image::{Image, ImageIO, ImageMetadata}; use crate::core::light::CreateLight; use crate::core::spectrum::spectrum_to_photometric; use crate::spectra::colorspace::new; -use crate::utils::{Ptr, Upload, resolve_filename}; -use shared::PI; +use crate::utils::{Arena, ParameterDictionary, Ptr, Upload, resolve_filename}; use shared::core::geometry::{Bounds2f, Frame, VectorLike}; use shared::core::image::ImageAccess; use shared::core::light::{Light, LightBase, LightType}; @@ -14,21 +13,28 @@ use shared::lights::{ProjectionLight, SpotLight}; use shared::spectra::RGBColorSpace; use shared::utils::math::{radians, square}; use shared::utils::{Ptr, Transform}; +use shared::{Float, PI}; -impl CreateLight for SpotLight { +pub trait CreateSpotLight { fn new( render_from_light: Transform, medium_interface: MediumInterface, le: Spectrum, scale: shared::Float, - shape: Option>, - alpha: Option>, - image: Option>, - image_color_space: Option>, - two_sided: Option, - fov: Option, - cos_falloff_start: Option, - total_width: Option, + cos_falloff_start: Float, + total_width: Float, + ) -> Self; +} + +impl CreateSpotLight for SpotLight { + fn new( + render_from_light: Transform, + medium_interface: MediumInterface, + le: Spectrum, + scale: shared::Float, + image: Ptr, + cos_falloff_start: Float, + total_width: Float, ) -> Self { let base = LightBase::new( LightType::DeltaPosition, @@ -41,21 +47,23 @@ impl CreateLight for SpotLight { base, iemit, scale, - cos_falloff_end: radians(total_width.unwrap().cos()), - cos_falloff_start: radians(cos_falloff_start.unwrap().cos()), + cos_falloff_end: radians(total_width).cos(), + cos_falloff_start: radians(cos_falloff_start).cos(), } } +} +impl CreateLight for SpotLight { fn create( - arena: &mut crate::utils::Arena, + arena: &mut Arena, render_from_light: Transform, medium: Medium, - parameters: &crate::utils::ParameterDictionary, + parameters: &ParameterDictionary, loc: &FileLoc, shape: &Shape, alpha_tex: &FloatTexture, colorspace: Option<&RGBColorSpace>, - ) -> Light { + ) -> Result { let i = parameters .get_one_spectrum( "I", @@ -87,14 +95,9 @@ impl CreateLight for SpotLight { medium.into(), le, scale, - None, - None, - None, - None, - None, - None, - Some(coneangle), - Some(coneangle - conedelta), + coneangle, + coneangle - conedelta, ); + Ok(Light::Spot(specific)) } } diff --git a/src/materials/dielectric.rs b/src/materials/dielectric.rs index cce0ded..b2a97fa 100644 --- a/src/materials/dielectric.rs +++ b/src/materials/dielectric.rs @@ -9,8 +9,8 @@ use crate::core::scattering::TrowbridgeReitzDistribution; use crate::core::spectrum::{Spectrum, SpectrumTrait}; use crate::core::texture::{GPUFloatTexture, GPUSpectrumTexture, TextureEvaluator}; use crate::spectra::{SampledSpectrum, SampledWavelengths}; -use crate::utils::Ptr; use crate::utils::math::clamp; +use shared::utils::Ptr; #[repr(C)] #[derive(Clone, Copy, Debug)] diff --git a/src/utils/arena.rs b/src/utils/arena.rs index bd838e3..a8b01d7 100644 --- a/src/utils/arena.rs +++ b/src/utils/arena.rs @@ -1,7 +1,6 @@ use crate::core::image::Image; use crate::core::texture::FloatTexture; use crate::utils::sampling::PiecewiseConstant2D; -use core::alloc::Layout; use shared::Float; use shared::core::color::RGBToSpectrumTable; use shared::core::image::DeviceImage; @@ -11,61 +10,75 @@ use shared::spectra::{RGBColorSpace, StandardColorSpaces}; use shared::textures::*; use shared::utils::Ptr; use shared::utils::sampling::DevicePiecewiseConstant2D; +use std::alloc::Layout; pub struct Arena { - buffer: Vec, + buffer: Vec<(*mut u8, Layout)>, } impl Arena { pub fn new() -> Self { - Self { - buffer: Vec::with_capacity(64 * 1024 * 1024), - } + Self { buffer: Vec::new() } } - pub fn alloc(&mut self, value: T) -> Ptr { + pub fn alloc(&mut self, value: T) -> Ptr { let layout = Layout::new::(); - let offset = self.alloc_raw(layout); + let ptr = unsafe { self.alloc_unified(layout) } as *mut T; + + // Write the value unsafe { - let ptr = self.buffer.as_mut_ptr().add(offset) as *mut T; - std::ptr::write(ptr, value); + ptr.write(value); } - Ptr { - offset: offset as i32, - _marker: std::marker::PhantomData, + Ptr::from_raw(ptr) + } + + pub fn alloc_opt(&mut self, value: Option) -> Ptr { + match value { + Some(v) => self.alloc(v), + None => Ptr::null(), } } - pub fn alloc_slice(&mut self, values: &[T]) -> Ptr { + pub fn alloc_slice(&mut self, values: &[T]) -> (Ptr, usize) { + if values.is_empty() { + return (Ptr::null(), 0); + } + let layout = Layout::array::(values.len()).unwrap(); - let offset = self.alloc_raw(layout); + let ptr = unsafe { self.alloc_unified(layout) } as *mut T; unsafe { - let ptr = self.buffer.as_mut_ptr().add(offset) as *mut T; std::ptr::copy_nonoverlapping(values.as_ptr(), ptr, values.len()); } - Ptr { - offset: offset as i32, - _marker: std::marker::PhantomData, - } + (Ptr::from_raw(ptr), values.len()) } - fn alloc_raw(&mut self, layout: Layout) -> usize { - let len = self.buffer.len(); - let align_offset = (len + layout.align() - 1) & !(layout.align() - 1); - let new_len = align_offset + layout.size(); + #[cfg(feature = "cuda")] + unsafe fn alloc_unified(&mut self, layout: Layout) -> *mut u8 { + use cuda_runtime_sys::*; - if new_len > self.buffer.capacity() { - // Growth strategy: Double capacity to reduce frequency of resizing - let new_cap = std::cmp::max(self.buffer.capacity() * 2, new_len); - self.buffer.reserve(new_cap - self.buffer.len()); + let mut ptr: *mut std::ffi::c_void = std::ptr::null_mut(); + let size = layout.size().max(layout.align()); + + let result = cudaMallocManaged(&mut ptr, size, cudaMemAttachGlobal); + + if result != cudaError::cudaSuccess { + panic!("cudaMallocManaged failed: {:?}", result); } - self.buffer.resize(new_len, 0); - align_offset + self.allocations.push((ptr as *mut u8, layout)); + ptr as *mut u8 + } + + #[cfg(not(feature = "cuda"))] + unsafe fn alloc_unified(&mut self, layout: Layout) -> *mut u8 { + // Fallback: regular allocation for CPU-only testing + let ptr = std::alloc::alloc(layout); + self.allocations.push((ptr, layout)); + ptr } pub fn raw_data(&self) -> &[u8] { @@ -73,6 +86,23 @@ impl Arena { } } +impl Drop for UnifiedArena { + fn drop(&mut self) { + for (ptr, layout) in self.allocations.drain(..) { + unsafe { + #[cfg(feature = "cuda")] + { + cuda_runtime_sys::cudaFree(ptr as *mut _); + } + #[cfg(not(feature = "cuda"))] + { + std::alloc::dealloc(ptr, layout); + } + } + } + } +} + pub trait Upload { type Target: Copy; @@ -87,10 +117,10 @@ impl Upload for Shape { } impl Upload for Image { - type Target = DeviceImage; + type Target = Image; fn upload(&self, arena: &mut Arena) -> Ptr { let pixels_ptr = arena.alloc_slice(&self.storage_as_slice()); - let device_img = DeviceImage { + let device_img = Image { base: self.base, pixels: pixels_ptr, }; diff --git a/src/utils/containers.rs b/src/utils/containers.rs index f085c25..3ee3a0f 100644 --- a/src/utils/containers.rs +++ b/src/utils/containers.rs @@ -1,5 +1,5 @@ use crate::core::geometry::{Bounds2i, Point2i}; -use crate::shared::utils::containers::Array2D; +use shared::utils::containers::Array2D; use std::ops::{Deref, DerefMut}; pub struct InternCache { @@ -20,7 +20,7 @@ where let mut lock = self.cache.lock().unwrap(); if let Some(existing) = lock.get(&value) { - return existing.clone(); // Returns a cheap Arc copy + return existing.clone(); } let new_item = Arc::new(value); diff --git a/src/utils/mod.rs b/src/utils/mod.rs index 65d51dd..79c4955 100644 --- a/src/utils/mod.rs +++ b/src/utils/mod.rs @@ -9,6 +9,7 @@ pub mod parallel; pub mod parameters; pub mod parser; pub mod sampling; +pub mod strings; pub use arena::{Arena, Upload}; pub use error::FileLoc; diff --git a/src/utils/sampling.rs b/src/utils/sampling.rs index b0e55b8..6bdf713 100644 --- a/src/utils/sampling.rs +++ b/src/utils/sampling.rs @@ -8,14 +8,13 @@ use std::sync::Arc; #[derive(Debug, Clone)] pub struct PiecewiseConstant1D { - pub func: Vec, - pub cdf: Vec, - pub func_integral: Float, - pub min: Float, - pub max: Float, + func: Box<[Float]>, + cdf: Box<[Float]>, + pub device: DevicePiecewiseConstant1D, } impl PiecewiseConstant1D { + // Constructors pub fn new(f: &[Float]) -> Self { Self::new_with_bounds(f, 0.0, 1.0) } @@ -34,83 +33,154 @@ impl PiecewiseConstant1D { } } - pub fn new_with_bounds(f: &[Float], min: Float, max: Float) -> Self { + pub fn from_func(f: F, min: Float, max: Float, n: usize) -> Self + where + F: Fn(Float) -> Float, + { + let delta = (max - min) / n as Float; + let values: Vec = (0..n) + .map(|i| { + let x = min + (i as Float + 0.5) * delta; + f(x) + }) + .collect(); + + Self::new_with_bounds(values, min, max) + } + + pub fn new_with_bounds(f: Vec, min: Float, max: Float) -> Self { let n = f.len(); + let mut cdf = Vec::with_capacity(n + 1); + cdf.push(0.0); - let mut func_vec = f.to_vec(); - let mut cdf_vec = vec![0.0; n + 1]; - - cdf_vec[0] = 0.0; - for i in 1..=n { - cdf_vec[i] = cdf_vec[i - 1] + func_vec[i - 1] / n as Float; + let delta = (max - min) / n as Float; + for i in 0..n { + cdf.push(cdf[i] + f[i] * delta); } - let func_integral = cdf_vec[n]; + let func_integral = cdf[n]; + if func_integral > 0.0 { - for i in 1..=n { - cdf_vec[i] /= func_integral; - } - } else { - for i in 1..=n { - cdf_vec[i] = i as Float / n as Float; + for c in &mut cdf { + *c /= func_integral; } } - Self { - func: func_vec, - cdf: cdf_vec, - func_integral, + // Convert to boxed slices (no more reallocation possible) + let func: Box<[Float]> = f.into_boxed_slice(); + let cdf: Box<[Float]> = cdf.into_boxed_slice(); + + let device = DevicePiecewiseConstant1D { + func: func.as_ptr(), + cdf: cdf.as_ptr(), min, max, - } + n: n as u32, + func_integral, + }; + + Self { func, cdf, device } + } + + // Accessors + pub fn min(&self) -> Float { + self.device.min + } + pub fn max(&self) -> Float { + self.device.max + } + pub fn n(&self) -> usize { + self.device.n as usize + } + pub fn integral(&self) -> Float { + self.device.func_integral + } + + pub fn func(&self) -> &[Float] { + &self.func + } + pub fn cdf(&self) -> &[Float] { + &self.cdf + } +} + +impl std::ops::Deref for PiecewiseConstant1D { + type Target = DevicePiecewiseConstant1D; + + fn deref(&self) -> &Self::Target { + &self.device } } -#[derive(Debug, Clone)] pub struct PiecewiseConstant2D { - pub domain: Bounds2f, - pub p_marginal: PiecewiseConstant1D, - pub p_conditionals: Vec, + conditionals: Vec, + marginal: PiecewiseConstant1D, + conditional_devices: Box<[DevicePiecewiseConstant1D]>, + pub device: DevicePiecewiseConstant2D, } impl PiecewiseConstant2D { - pub fn new(data: &Array2D, x_size: u32, y_size: u32, domain: Bounds2f) -> Self { - let nu = x_size as usize; - let nv = y_size as usize; - let mut conditionals = Vec::with_capacity(nv); - for v in 0..nv { - let row = unsafe { core::slice::from_raw_parts(data.values.add(v * nu), nu) }; - conditionals.push(DevicePiecewiseConstant1D::new_with_bounds( - row, - domain.p_min.x(), - domain.p_max.x(), - )); + pub fn new(data: &[Float], n_u: usize, n_v: usize) -> Self { + assert_eq!(data.len(), n_u * n_v); + + // Build conditional distributions p(u|v) for each row + let mut conditionals = Vec::with_capacity(n_v); + let mut marginal_func = Vec::with_capacity(n_v); + + for v in 0..n_v { + let row_start = v * n_u; + let row: Vec = data[row_start..row_start + n_u].to_vec(); + + let conditional = PiecewiseConstant1D::new_with_bounds(row, 0.0, 1.0); + marginal_func.push(conditional.integral()); + conditionals.push(conditional); } - let marginal_funcs: Vec = conditionals.iter().map(|c| c.func_integral).collect(); - let p_marginal = DevicePiecewiseConstant1D::new_with_bounds( - &marginal_funcs, - domain.p_min.y(), - domain.p_max.y(), - ); + // Build marginal distribution p(v) + let marginal = PiecewiseConstant1D::new_with_bounds(marginal_func, 0.0, 1.0); + + // Create array of device structs + let conditional_devices: Box<[DevicePiecewiseConstant1D]> = conditionals + .iter() + .map(|c| c.device) + .collect::>() + .into_boxed_slice(); + + let device = DevicePiecewiseConstant2D { + conditional: conditional_devices.as_ptr(), + marginal: marginal.device, + n_u: n_u as u32, + n_v: n_v as u32, + }; Self { - domain, - p_marginal, - p_conditionals: conditionals, + conditionals, + marginal, + conditional_devices, + device, } } - pub fn new_with_bounds(data: &Array2D, domain: Bounds2f) -> Self { - Self::new(data, data.x_size(), data.y_size(), domain) + pub fn from_image(image: &Image) -> Self { + let res = image.resolution(); + let n_u = res.x() as usize; + let n_v = res.y() as usize; + + let mut data = Vec::with_capacity(n_u * n_v); + + for v in 0..n_v { + for u in 0..n_u { + let p = Point2i::new(u as i32, v as i32); + let luminance = image.get_channels(p).average(); + data.push(luminance); + } + } + + Self::new(&data, n_u, n_v) } - pub fn new_with_data(data: &Array2D) -> Self { - let nx = data.x_size(); - let ny = data.y_size(); - let domain = Bounds2f::from_points(Point2f::new(0.0, 0.0), Point2f::new(1.0, 1.0)); - - Self::new(data, nx, ny, domain) + pub fn integral(&self) -> Float { + self.marginal.integral() } }