From 86c9a90f2e075aba40d67d651d5e99bd12c76e0c Mon Sep 17 00:00:00 2001 From: pingu Date: Sun, 18 Jan 2026 16:29:27 +0000 Subject: [PATCH] Fixes --- Cargo.toml | 8 +- kernels/src/lib.rs | 451 +---------------------- kernels/src/wavefront/mod.rs | 47 --- kernels/src/workitem.rs | 535 ---------------------------- shared/build.rs | 5 + shared/src/bxdfs/layered.rs | 9 +- shared/src/bxdfs/measured.rs | 8 +- shared/src/cameras/mod.rs | 2 +- shared/src/cameras/realistic.rs | 69 ++-- shared/src/cameras/spherical.rs | 2 +- shared/src/core/bsdf.rs | 13 +- shared/src/core/bssrdf.rs | 19 +- shared/src/core/camera.rs | 83 ++--- shared/src/core/color.rs | 66 +++- shared/src/core/film.rs | 213 +++-------- shared/src/core/filter.rs | 30 +- shared/src/core/geometry/ray.rs | 12 +- shared/src/core/image.rs | 10 + shared/src/core/interaction.rs | 35 +- shared/src/core/light.rs | 1 - shared/src/core/material.rs | 6 +- shared/src/core/medium.rs | 239 ++++--------- shared/src/core/primitive.rs | 41 ++- shared/src/core/sampler.rs | 122 +++---- shared/src/core/shape.rs | 19 +- shared/src/core/spectrum.rs | 3 +- shared/src/core/texture.rs | 14 +- shared/src/filters/gaussian.rs | 21 -- shared/src/filters/lanczos.rs | 14 - shared/src/filters/mitchell.rs | 17 +- shared/src/lights/diffuse.rs | 14 +- shared/src/lights/infinite.rs | 4 +- shared/src/lights/projection.rs | 13 +- shared/src/lights/sampler.rs | 36 +- shared/src/materials/coated.rs | 113 +++--- shared/src/materials/complex.rs | 64 ++-- shared/src/materials/dielectric.rs | 24 +- shared/src/materials/diffuse.rs | 22 +- shared/src/materials/mix.rs | 6 +- shared/src/shapes/bilinear.rs | 30 +- shared/src/shapes/triangle.rs | 32 +- shared/src/spectra/colorspace.rs | 14 +- shared/src/spectra/rgb.rs | 6 +- shared/src/spectra/simple.rs | 24 +- shared/src/textures/bilerp.rs | 18 +- shared/src/textures/checkerboard.rs | 14 +- shared/src/textures/dots.rs | 10 +- shared/src/textures/image.rs | 18 +- shared/src/textures/marble.rs | 4 +- shared/src/textures/ptex.rs | 10 +- shared/src/utils/containers.rs | 14 +- shared/src/utils/error.rs | 8 +- shared/src/utils/math.rs | 9 +- shared/src/utils/mod.rs | 15 +- shared/src/utils/ptr.rs | 9 +- shared/src/utils/sampling.rs | 35 +- src/core/aggregates.rs | 11 +- src/core/camera.rs | 1 + src/core/film.rs | 143 +++++++- src/core/filter.rs | 36 ++ src/core/image/io.rs | 57 ++- src/core/image/metadata.rs | 12 +- src/core/image/mod.rs | 13 +- src/core/image/ops.rs | 32 +- src/core/light.rs | 10 +- src/core/material.rs | 8 +- src/core/mod.rs | 1 + src/core/sampler.rs | 29 +- src/core/scene.rs | 19 +- src/core/shape.rs | 7 +- src/core/spectrum.rs | 1 - src/core/texture.rs | 29 +- src/integrators/mod.rs | 51 ++- src/integrators/pipeline.rs | 26 +- src/lib.rs | 4 + src/lights/diffuse.rs | 21 +- src/lights/distant.rs | 6 +- src/lights/goniometric.rs | 5 +- src/lights/infinite.rs | 3 +- src/lights/point.rs | 1 - src/lights/projection.rs | 5 +- src/lights/spot.rs | 16 +- src/materials/coated.rs | 48 +-- src/materials/complex.rs | 72 ++-- src/materials/conductor.rs | 19 +- src/materials/dielectric.rs | 17 +- src/materials/diffuse.rs | 18 +- src/materials/mix.rs | 19 +- src/samplers/halton.rs | 11 +- src/samplers/independent.rs | 11 +- src/samplers/sobol.rs | 34 +- src/samplers/stratified.rs | 15 +- src/shapes/bilinear.rs | 5 +- src/shapes/curves.rs | 7 +- src/shapes/mesh.rs | 27 +- src/shapes/mod.rs | 2 +- src/shapes/triangle.rs | 3 + src/spectra/colorspace.rs | 81 +++-- src/spectra/data.rs | 15 +- src/spectra/dense.rs | 8 +- src/spectra/mod.rs | 29 +- src/textures/image.rs | 17 +- src/textures/mix.rs | 6 +- src/textures/ptex.rs | 43 ++- src/textures/scaled.rs | 8 +- src/utils/arena.rs | 102 +++++- src/utils/containers.rs | 7 +- src/utils/error.rs | 4 +- src/utils/file.rs | 2 +- src/utils/math.rs | 8 +- src/utils/mipmap.rs | 9 +- src/utils/parameters.rs | 52 +-- src/utils/parser.rs | 3 - src/utils/sampling.rs | 40 ++- 114 files changed, 1548 insertions(+), 2421 deletions(-) delete mode 100644 kernels/src/wavefront/mod.rs delete mode 100644 kernels/src/workitem.rs create mode 100644 shared/build.rs diff --git a/Cargo.toml b/Cargo.toml index e3f9352..5aa6018 100644 --- a/Cargo.toml +++ b/Cargo.toml @@ -6,13 +6,14 @@ edition = "2024" [features] default = [] use_f64 = [] +use_gpu = [] +use_nvtx = [] cuda = ["cust", "cuda_builder", "shared/cuda", ] [dependencies] anyhow = "1.0.100" exr = "1.73.0" flate2 = "1.1.5" -gpu = "0.2.3" half = "2.7.1" image_rs = { package = "image", version = "0.25.8" } indicatif = "0.18.3" @@ -40,6 +41,11 @@ slice = "0.0.4" crossbeam-channel = "0.5.15" num_cpus = "1.17.0" ply-rs = "0.1.3" +enum_dispatch = "0.3.13" +bytemuck = "1.24.0" +once_cell = "1.21.3" +smallvec = "1.15.1" +cuda-runtime-sys = "0.3.0-alpha.1" [build-dependencies] spirv-builder = { git = "https://github.com/rust-gpu/rust-gpu", branch = "main", optional = true } diff --git a/kernels/src/lib.rs b/kernels/src/lib.rs index 3ed8d63..7a1317c 100644 --- a/kernels/src/lib.rs +++ b/kernels/src/lib.rs @@ -1,447 +1,16 @@ +#![no_std] + use cuda_std::prelude::*; +use shared::Float; -pub mod wavefront; -pub mod workitem; +#[kernel] +pub unsafe fn scale_array_kernel(data: *mut Float, n: u32, scale: Float) { + let i = thread::index_1d() as usize; -use cust::context::{CacheConfig, CurrentContext, ResourceLimit}; -use cust::device::DeviceAttribute; -use cust::memory::{DeviceCopy, DeviceMemory}; -use cust::prelude::*; -use lazy_static::lazy_static; -use parking_lot::Mutex; -use std::error::Error; -use std::ffi::c_void; -use std::sync::Arc; - -use crate::Float; -use crate::core::geometry::{Normal, Point, Vector}; -use crate::core::medium::Medium; -use crate::core::options::{PBRTOptions, get_options}; -use crate::impl_gpu_traits; -use crate::impl_math_gpu_traits; -use crate::spectra::{SampledSpectrum, SampledWavelengths}; -use crate::utils::interval::Interval; - -pub use workitem::{ - EscapedRayQueue, GetBSSRDFAndProbeRayQueue, HitAreaLightQueue, MaterialEvalQueue, - MediumSampleQueue, MediumScatterQueue, PixelSampleStateStorage, RayQueue, ShadowRayQueue, - SubsurfaceScatterQueue, -}; - -#[repr(C, align(16))] -#[derive(Clone, Copy, Debug, Default, PartialEq)] -pub struct Float4 { - pub v: [f32; 4], -} - -pub type Vec4 = Vector; - -impl From for Float4 { - #[inline] - fn from(vec: Vector) -> Self { - Self { v: vec.0 } - } -} - -impl From for Vec4 { - #[inline] - fn from(storage: Float4) -> Self { - Vector(storage.v) - } -} - -impl_math_gpu_traits!(Vector); -impl_math_gpu_traits!(Normal); -impl_math_gpu_traits!(Point); -impl_gpu_traits!(Interval); -impl_gpu_traits!(Float4); -impl_gpu_traits!(SampledSpectrum); -impl_gpu_traits!(SampledWavelengths); - -struct KernelStats { - description: String, - num_launches: usize, - sum_ms: f32, - min_ms: f32, - max_ms: f32, -} - -impl KernelStats { - fn new(description: &str) -> Self { - Self { - description: description.to_string(), - num_launches: 0, - sum_ms: 0.0, - min_ms: 0.0, - max_ms: 0.0, + if i < n as usize { + let ptr = unsafe { data.add(i) }; + unsafe { + *ptr = *ptr * scale; } } } - -struct ProfilerEvent { - start: Event, - stop: Event, - active: bool, - stats: Option>>, -} - -impl ProfilerEvent { - fn new() -> Result { - let start = Event::new(EventFlags::DEFAULT)?; - let stop = Event::new(EventFlags::DEFAULT)?; - Ok(Self { - start, - stop, - active: false, - stats: None, - }) - } - - fn sync(&mut self) { - if !self.active { - return; - } - - if self.stop.synchronize().is_ok() { - // Check timing between start and stop - match self.stop.elapsed_time_f32(&self.start) { - Ok(ms) => { - if let Some(stats_arc) = &self.stats { - let mut stats = stats_arc.lock(); - stats.num_launches += 1; - if stats.num_launches == 1 { - stats.sum_ms = ms; - stats.min_ms = ms; - stats.max_ms = ms; - } else { - stats.sum_ms += ms; - stats.min_ms = stats.min_ms.min(ms); - stats.max_ms = stats.max_ms.max(ms); - } - } - } - Err(e) => log::error!("Failed to get elapsed time: {:?}", e), - } - } - self.active = false; - } -} - -// --- Profiler Manager --- - -struct Profiler { - kernel_stats: Vec>>, - event_pool: Vec, - pool_offset: usize, -} - -impl Profiler { - fn new() -> Self { - Self { - kernel_stats: Vec::new(), - event_pool: Vec::new(), - pool_offset: 0, - } - } - - /// Prepares an event from the pool. - /// Returns a mutable reference to the event, valid as long as the borrow of self. - fn prepare<'a>(&'a mut self, description: &str) -> &'a mut ProfilerEvent { - // Grow pool if empty or needed (simple heuristic) - if self.event_pool.is_empty() { - for _ in 0..128 { - if let Ok(e) = ProfilerEvent::new() { - self.event_pool.push(e); - } - } - } - - if self.pool_offset >= self.event_pool.len() { - self.pool_offset = 0; - } - - let idx = self.pool_offset; - self.pool_offset += 1; - - let pe = &mut self.event_pool[idx]; - - if pe.active { - pe.sync(); - } - - pe.active = true; - pe.stats = None; - - // Find or create stats - let mut found = None; - for s in &self.kernel_stats { - if s.lock().description == description { - found = Some(s.clone()); - break; - } - } - - if found.is_none() { - let new_stats = Arc::new(Mutex::new(KernelStats::new(description))); - self.kernel_stats.push(new_stats.clone()); - found = Some(new_stats); - } - - pe.stats = found; - pe - } -} - -pub struct GpuState { - context: Context, - stream: Stream, - profiler: Profiler, -} - -impl GpuState { - fn init(device_index: u32) -> Result> { - cust::init(CudaFlags::empty())?; - - let device = Device::get_device(device_index)?; - - let name = device.name().unwrap_or_else(|_| "Unknown".into()); - let memory = device.total_memory().unwrap_or(0); - let memory_gb = memory as f64 / (1024.0 * 1024.0 * 1024.0); - - let major = device - .get_attribute(DeviceAttribute::ComputeCapabilityMajor) - .unwrap_or(0); - let minor = device - .get_attribute(DeviceAttribute::ComputeCapabilityMinor) - .unwrap_or(0); - - log::info!( - "Selected GPU: {} ({:.2} GB, SM {}.{})", - name, - memory_gb, - major, - minor - ); - - let has_unified = device - .get_attribute(DeviceAttribute::UnifiedAddressing) - .unwrap_or(0); - if has_unified == 0 { - panic!("Selected GPU does not support unified addressing."); - } - - let context = Context::new(device)?; - - CurrentContext::set_resource_limit(ResourceLimit::StackSize, 8192)?; - let stack_size = CurrentContext::get_resource_limit(ResourceLimit::StackSize)?; - log::info!("Reset stack size to {}", stack_size); - - CurrentContext::set_resource_limit(ResourceLimit::PrintfFifoSize, 32 * 1024 * 1024)?; - CurrentContext::set_cache_config(CacheConfig::PreferL1)?; - - let stream = Stream::new(StreamFlags::DEFAULT, None)?; - - Ok(Self { - context, - stream, - profiler: Profiler::new(), - }) - } -} - -lazy_static! { - static ref GPU_STATE: Mutex> = Mutex::new(None); -} - -pub fn gpu_init() { - if !get_options().use_gpu { - return; - } - - let device_id = get_options().gpu_device.unwrap_or(0); - log::info!("Initializing GPU Device {}", device_id); - - match GpuState::init(device_id) { - Ok(state) => { - #[cfg(feature = "use_nvtx")] - nvtx::name_thread("MAIN_THREAD"); - *GPU_STATE.lock() = Some(state); - } - Err(e) => { - panic!("Failed to initialize GPU: {:?}", e); - } - } -} - -pub fn gpu_thread_init() { - if let Some(state) = GPU_STATE.lock().as_ref() { - if let Err(e) = CurrentContext::set_current(&state.context) { - log::error!("Failed to set CUDA context for thread: {:?}", e); - } - } -} - -pub fn gpu_wait() { - let mut guard = GPU_STATE.lock(); - if let Some(state) = guard.as_mut() { - if let Err(e) = state.stream.synchronize() { - log::error!("GPU Wait failed: {:?}", e); - } - } -} - -/// Launches a parallel for loop on the GPU. -/// -/// # Arguments -/// * `description`: Name for profiling. -/// * `n_items`: Total items (threads). -/// * `function`: Compiled kernel function handle. -/// * `params`: Kernel parameters (must be DeviceCopy). -pub fn gpu_parallel_for( - description: &str, - n_items: i32, - function: &Function, - params: &T, -) { - #[cfg(feature = "use_nvtx")] - nvtx::range_push(description); - - let mut guard = GPU_STATE.lock(); - let state = guard.as_mut().expect("GPU not initialized"); - - let (_, block_size) = match function.suggested_launch_configuration(0, 0.into()) { - Ok(cfg) => cfg, - Err(e) => panic!( - "Failed to calculate launch config for {}: {:?}", - description, e - ), - }; - - #[cfg(debug_assertions)] - log::debug!("[{}] Block size: {}", description, block_size); - - let grid_size = (n_items as u32 + block_size - 1) / block_size; - - let stream = &state.stream; - let profiler = &mut state.profiler; - - // Save the index we are about to use so we can retrieve the STOP event later - let event_idx = profiler.pool_offset; - - { - let pe = profiler.prepare(description); - if let Err(e) = pe.start.record(stream) { - log::error!("Failed to record start event: {:?}", e); - } - } - - let params_ptr = params as *const T as *mut c_void; - let n_items_ptr = &n_items as *const i32 as *mut c_void; - let args = [params_ptr, n_items_ptr]; - - unsafe { - if let Err(e) = - state - .stream - .launch(function, (grid_size, 1, 1), (block_size, 1, 1), 0, &args) - { - panic!("CUDA Launch failed for {}: {:?}", description, e); - } - } - - // Retrieve the specific event we just set up. - // Pool_offset was incremented in prepare(). - // If event_idx was the one used, the event is at event_idx. - if event_idx < profiler.event_pool.len() { - let pe = &mut profiler.event_pool[event_idx]; - if let Err(e) = pe.stop.record(stream) { - log::error!("Failed to record stop event: {:?}", e); - } - } - - #[cfg(debug_assertions)] - let _ = state.stream.synchronize(); - - #[cfg(feature = "use_nvtx")] - nvtx::range_pop(); -} - -pub fn report_kernel_stats() { - let mut guard = GPU_STATE.lock(); - if let Some(state) = guard.as_mut() { - let _ = state.stream.synchronize(); - - // Process all pending events - for pe in &mut state.profiler.event_pool { - if pe.active { - pe.sync(); - } - } - - let mut total_ms = 0.0; - for s in &state.profiler.kernel_stats { - total_ms += s.lock().sum_ms; - } - - println!("Wavefront Kernel Profile:"); - for s in &state.profiler.kernel_stats { - let stats = s.lock(); - let percent = if total_ms > 0.0 { - 100.0 * stats.sum_ms / total_ms - } else { - 0.0 - }; - println!( - " {:<45} {:5} launches {:9.2} ms / {:5.1}% (avg {:6.3})", - stats.description, - stats.num_launches, - stats.sum_ms, - percent, - if stats.num_launches > 0 { - stats.sum_ms / stats.num_launches as f32 - } else { - 0.0 - } - ); - } - println!("\nTotal: {:.2} ms", total_ms); - } -} - -pub fn gpu_memset(dst: &mut DeviceSlice, value: u8) { - unsafe { - let ptr = dst.as_raw_ptr(); // Returns CUdeviceptr (u64) - let len = dst.len() * std::mem::size_of::(); - - // We need the `cust::external::cuda` or equivalent sys crate function - - log::warn!("gpu_memset requested but raw memset not exposed via safe cust API yet."); - } -} - -#[macro_export] -macro_rules! impl_gpu_traits { - ($name:ty) => { - unsafe impl cust::memory::DeviceCopy for $name {} - unsafe impl bytemuck::Zeroable for $name {} - unsafe impl bytemuck::Pod for $name {} - }; -} - -#[macro_export] -macro_rules! impl_math_gpu_traits { - ($Struct:ident) => { - #[cfg(feature = "use_gpu")] - unsafe impl cust::memory::DeviceCopy for $Struct where - T: cust::memory::DeviceCopy + Copy - { - } - - unsafe impl bytemuck::Zeroable for $Struct where - T: bytemuck::Zeroable - { - } - - unsafe impl bytemuck::Pod for $Struct where T: bytemuck::Pod {} - }; -} diff --git a/kernels/src/wavefront/mod.rs b/kernels/src/wavefront/mod.rs deleted file mode 100644 index cbba9a1..0000000 --- a/kernels/src/wavefront/mod.rs +++ /dev/null @@ -1,47 +0,0 @@ -use image_rs::Pixel; - -use crate::camera::Camera; -use crate::core::film::Film; -use crate::core::filter::Filter; -use crate::core::sampler::Sampler; -use crate::core::scene::BasicScene; -use crate::lights::Light; -use crate::lights::LightSampler; -use crate::{ - EscapedRayQueue, GetBSSRDFAndProbeRayQueue, HitAreaLightQueue, MaterialEvalQueue, - MediumSampleQueue, MediumScatterQueue, PixelSampleStateStorage, RayQueue, ShadowRayQueue, - SubsurfaceScatterQueue, -}; -use std::sync::Arc; - -pub struct WavefrontPathIntegrator { - pub film: Film, - pub filter: Filter, - pub sampler: Sampler, - pub camera: Arc, - pub light_sampler: LightSampler, - pub infinite_lights: Option>>, - pub max_depth: i32, - pub samples_per_pixel: i32, - pub regularize: bool, - pub scanlines_per_pixel: i32, - pub max_queue_size: i32, - pub pixel_sample_state: PixelSampleStateStorage, - pub ray_queue: [RayQueue; 2], - pub hit_area_light_queue: HitAreaLightQueue, - pub shadow_ray_queue: ShadowRayQueue, - pub escaped_ray_queue: Option, - pub basic_material_queue: Option, - pub universal_material_queue: Option, - pub medium_sample_queue: Option, - pub medium_scatter_queue: Option, - pub bssrf_queue: Option, - pub subsurface_queue: Option, -} - -#[cfg(feature = "use_gpu")] -impl WavefrontPathIntegrator { - pub fn new(scene: BasicScene) -> Self { - todo!() - } -} diff --git a/kernels/src/workitem.rs b/kernels/src/workitem.rs deleted file mode 100644 index f85ac43..0000000 --- a/kernels/src/workitem.rs +++ /dev/null @@ -1,535 +0,0 @@ -#![allow(clippy::too_many_arguments)] -use super::Float4; -use crate::Float; -use crate::core::geometry::{Normal3f, Point2f, Point2i, Point3f, Point3fi, Ray, Vector3f}; -use crate::lights::LightSampleContext; -use crate::soa_struct; -use crate::spectra::{SampledSpectrum, SampledWavelengths}; -use cust::memory::{CopyDestination, DeviceMemory}; -use cust::prelude::*; - -#[macro_export] -macro_rules! soa_struct { - ( - $(#[$outer:meta])* - pub struct $name:ident { - $( - pub $field:ident : $type:ty - ),* $(,)? - } -) => { - #[cfg(feature = "use_gpu")] - $(#[$outer])* - pub struct $name { - capacity: u32, - pub count: cust::memory::DeviceBuffer, - $( - pub $field: cust::memory::DeviceBuffer<$type>, - )* - } - - #[cfg(feature = "use_gpu")] - impl $name { - pub fn new(capacity: usize) -> cust::error::CudaResult { - use cust::memory::DeviceBuffer; - Ok(Self { - capacity: capacity as u32, - count: DeviceBuffer::zeroed(1)?, - $( - $field: DeviceBuffer::zeroed(capacity)?, - )* - }) - } - - pub fn len(&self) -> cust::error::CudaResult { - let mut host_count = [0u32; 1]; - self.count.copy_to(&mut host_count)?; - Ok(host_count[0]) - } - - pub fn reset(&mut self) -> cust::error::CudaResult<()> { - self.count.copy_from(&[0]) - } - - // Generate the View name - pub fn as_view(&mut self) -> paste::paste! { [<$name View>] } { - paste::paste! { - [<$name View>] { - capacity: self.capacity, - count: self.count.as_device_ptr().as_mut_ptr(), - $( - $field: self.$field.as_device_ptr().as_raw() as *mut $type, - )* - } - } - } - } - - paste::paste! { - #[repr(C)] - #[derive(Clone, Copy)] - pub struct [<$name View>] { - pub capacity: u32, - pub count: *mut u32, - $( - pub $field: *mut $type, - )* - } - - unsafe impl cust::memory::DeviceCopy for [<$name View>] {} - - impl [<$name View>] { - // The raw push that fills every field - #[cfg(feature = "use_gpu")] - pub unsafe fn push(&self, $( $field : $type ),* ) -> Option { - use core::sync::atomic::{AtomicU32, Ordering}; - - let index = unsafe { - let counter_ptr = self.count as *mut AtomicU32; - (*counter_ptr).fetch_add(1, Ordering::Relaxed) - }; - - if index >= self.capacity { - return None; - } - - unsafe { - $( - *self.$field.add(index as usize) = $field; - )* - } - - Some(index) - } - - #[cfg(feature = "use_gpu")] - pub unsafe fn size(&self) -> u32 { - use core::sync::atomic::{AtomicU32, Ordering}; - unsafe { - (*(self.count as *const AtomicU32)).load(Ordering::Relaxed) - } - } - - $( - #[cfg(feature = "use_gpu")] - pub fn [<$field _ptr>](&self) -> *mut $type { - self.$field - } - )* - } - } - }; -} - -#[repr(C)] -#[derive(Clone, Copy, Default)] -pub struct RaySamplesDirect { - pub u: Point2f, - pub uc: Float, -} - -#[repr(C)] -#[derive(Clone, Copy, Default)] -pub struct RaySamplesIndirect { - pub uc: Float, - pub rr: Float, - pub u: Point2f, -} - -#[repr(C)] -#[derive(Clone, Copy, Default)] -pub struct RaySamplesSubsurface { - pub uc: Float, - pub u: Point2f, -} - -#[repr(C)] -#[derive(Clone, Copy, Default)] -pub struct RaySamples { - pub direct: RaySamplesDirect, - pub indirect: RaySamplesIndirect, - pub have_subsurface: bool, - pub subsurface: RaySamplesSubsurface, -} - -soa_struct! { - pub struct RayQueue { - pub ray_o: Point3f, - pub ray_d: Vector3f, - - pub depth: i32, - pub lambda: SampledWavelengths, - pub pixel_index: u32, - - pub beta: SampledSpectrum, - pub r_u: SampledSpectrum, - pub r_l: SampledSpectrum, - - pub ctx_pi: Point3f, - pub ctx_n: Normal3f, - pub ctx_ns: Normal3f, - - pub eta_scale: Float, - pub specular_bounce: u32, - pub any_non_specular_bounces: u32, - } -} - -soa_struct! { - pub struct PixelSampleStateStorage { - pub p_pixel: Point2i, - pub l: SampledSpectrum, - pub lambda: SampledWavelengths, - pub filter_weight: Float, - pub visible_surface: u32, - pub camera_ray_weight: SampledSpectrum, - - pub rs_direct_packed: Float4, - pub rs_indirect_packed: Float4, - pub rs_subsurface_packed: Float4, - } -} - -soa_struct! { - pub struct EscapedRayQueue { - pub ray_o: Point3f, - pub ray_d: Vector3f, - pub depth: i32, - pub lambda: SampledWavelengths, - pub pixel_index: u32, - pub beta: SampledSpectrum, - pub specular_bounce: u32, - pub r_u: SampledSpectrum, - pub r_l: SampledSpectrum, - pub ctx_pi: Point3f, - pub ctx_n: Normal3f, - pub ctx_ns: Normal3f, - } -} - -soa_struct! { - pub struct HitAreaLightQueue { - pub area_light_id: u32, // Light ID - pub p: Point3f, - pub n: Normal3f, - pub uv: Point2f, - pub wo: Vector3f, - pub lambda: SampledWavelengths, - pub depth: i32, - pub beta: SampledSpectrum, - pub r_u: SampledSpectrum, - pub r_l: SampledSpectrum, - pub ctx_pi: Point3f, - pub ctx_n: Normal3f, - pub ctx_ns: Normal3f, - pub specular_bounce: u32, - pub pixel_index: u32, - } -} - -soa_struct! { - pub struct ShadowRayQueue { - pub ray_o: Point3f, - pub ray_d: Vector3f, - pub t_max: Float, - pub lambda: SampledWavelengths, - pub ld: SampledSpectrum, - pub r_u: SampledSpectrum, - pub r_l: SampledSpectrum, - pub pixel_index: u32, - } -} - -soa_struct! { - pub struct GetBSSRDFAndProbeRayQueue { - pub material_id: u32, - pub lambda: SampledWavelengths, - pub beta: SampledSpectrum, - pub r_u: SampledSpectrum, - pub p: Point3f, - pub wo: Vector3f, - pub n: Normal3f, - pub ns: Normal3f, - pub dpdus: Vector3f, - pub uv: Point2f, - pub depth: i32, - pub mi_inside: u32, - pub mi_outside: u32, - pub eta_scale: Float, - pub pixel_index: u32, - } -} - -soa_struct! { - pub struct SubsurfaceScatterQueue { - pub p0: Point3f, - pub p1: Point3f, - pub depth: i32, - pub material_id: u32, - pub lambda: SampledWavelengths, - pub beta: SampledSpectrum, - pub r_u: SampledSpectrum, - pub mi_inside: u32, - pub mi_outside: u32, - pub eta_scale: Float, - pub pixel_index: u32, - } -} - -soa_struct! { - pub struct MediumSampleQueue { - pub ray_o: Point3f, - pub ray_d: Vector3f, - pub t_max: Float, - pub lambda: SampledWavelengths, - pub beta: SampledSpectrum, - pub r_u: SampledSpectrum, - pub r_l: SampledSpectrum, - pub pixel_index: u32, - - pub ctx_pi: Point3f, - pub ctx_n: Normal3f, - pub ctx_ns: Normal3f, - - pub specular_bounce: u32, - pub any_non_specular_bounces: u32, - pub eta_scale: Float, - - pub area_light_id: u32, - pub pi: Point3fi, - pub n: Normal3f, - pub dpdu: Vector3f, - pub dpdv: Vector3f, - pub wo: Vector3f, - pub uv: Point2f, - pub material_id: u32, - pub ns: Normal3f, - pub dpdus: Vector3f, - pub dpdvs: Vector3f, - pub dndus: Normal3f, - pub dndvs: Normal3f, - pub face_index: i32, - pub mi_inside: u32, - pub mi_outside: u32, - } -} - -soa_struct! { - pub struct MaterialEvalQueue { - pub material_id: u32, - pub pi: Point3fi, - pub n: Normal3f, - pub dpdu: Vector3f, - pub dpdv: Vector3f, - pub time: Float, - pub depth: i32, - pub ns: Normal3f, - pub dpdus: Vector3f, - pub dpdvs: Vector3f, - pub dndus: Normal3f, - pub dndvs: Normal3f, - pub uv: Point2f, - pub face_index: i32, - pub lambda: SampledWavelengths, - pub pixel_index: u32, - pub any_non_specular_bounces: u32, - pub wo: Vector3f, - pub beta: SampledSpectrum, - pub r_u: SampledSpectrum, - pub eta_scale: Float, - pub mi_inside: u32, - pub mi_outside: u32, - } -} - -soa_struct! { - pub struct MediumScatterQueue { - pub p: Point3f, - pub depth: usize, - pub lambda: SampledWavelengths, - pub beta: SampledSpectrum, - pub r_u: SampledSpectrum, - pub wo: Vector3f, - pub time: Float, - pub eta_scale: Float, - pub pixel_index: usize, - - // ID - pub phase_function: u32, - pub medium: u32, - } -} - -#[repr(C)] -#[derive(Clone, Copy)] -pub struct RayWorkItem { - pub ray: Ray, - pub depth: i32, - pub lambda: SampledWavelengths, - pub pixel_index: u32, - pub beta: SampledSpectrum, - pub r_u: SampledSpectrum, - pub r_l: SampledSpectrum, - pub prev_intr_ctx: LightSampleContext, - pub eta_scale: Float, - pub specular_bounce: bool, - pub any_non_specular_bounces: bool, -} - -#[repr(C)] -#[derive(Clone, Copy)] -pub struct EscapedRayWorkItem { - pub ray_o: Point3f, - pub ray_d: Vector3f, - pub depth: i32, - pub lambda: SampledWavelengths, - pub pixel_index: u32, - pub beta: SampledSpectrum, - pub specular_bounce: bool, - pub r_u: SampledSpectrum, - pub r_l: SampledSpectrum, - pub prev_intr_ctx: LightSampleContext, -} - -#[repr(C)] -#[derive(Clone, Copy)] -pub struct ShadowRayWorkItem { - pub ray: Ray, - pub t_max: Float, - pub lambda: SampledWavelengths, - pub ld: SampledSpectrum, - pub r_u: SampledSpectrum, - pub r_l: SampledSpectrum, - pub pixel_index: u32, -} - -impl RayQueueView { - #[cfg(feature = "use_gpu")] - pub unsafe fn push_work_item(&self, item: RayWorkItem) -> Option { - unsafe { - self.push( - item.ray.o, - item.ray.d, - item.depth, - item.lambda, - item.pixel_index, - item.beta, - item.r_u, - item.r_l, - item.prev_intr_ctx.pi.into(), - item.prev_intr_ctx.n, - item.prev_intr_ctx.ns, - item.eta_scale, - if item.specular_bounce { 1 } else { 0 }, - if item.any_non_specular_bounces { 1 } else { 0 }, - ) - } - } -} - -impl EscapedRayQueueView { - #[cfg(feature = "use_gpu")] - pub unsafe fn push_work_item(&self, r: &RayWorkItem) -> Option { - unsafe { - self.push( - r.ray.o, - r.ray.d, - r.depth, - r.lambda, - r.pixel_index, - r.beta, - if r.specular_bounce { 1 } else { 0 }, - r.r_u, - r.r_l, - r.prev_intr_ctx.pi.into(), - r.prev_intr_ctx.n, - r.prev_intr_ctx.ns, - ) - } - } -} - -impl PixelSampleStateStorageView { - #[cfg(feature = "use_gpu")] - pub unsafe fn get_samples(&self, index: u32) -> RaySamples { - let i = index as usize; - - let (dir, ind, ss) = unsafe { - ( - *self.rs_direct_packed.add(i), - *self.rs_indirect_packed.add(i), - *self.rs_subsurface_packed.add(i), - ) - }; - - let direct_u = Point2f::new(dir.v[0], dir.v[1]); - let direct_uc = dir.v[2]; - let flags = dir.v[3] as i32; - let have_subsurface = (flags & 1) != 0; - - let indirect_uc = ind.v[0]; - let indirect_rr = ind.v[1]; - let indirect_u = Point2f::new(ind.v[2], ind.v[3]); - - let subsurface_uc = ss.v[0]; - let subsurface_u = Point2f::new(ss.v[1], ss.v[2]); - - RaySamples { - direct: RaySamplesDirect { - u: direct_u, - uc: direct_uc, - }, - indirect: RaySamplesIndirect { - uc: indirect_uc, - rr: indirect_rr, - u: indirect_u, - }, - have_subsurface, - subsurface: RaySamplesSubsurface { - uc: subsurface_uc, - u: subsurface_u, - }, - } - } - - #[cfg(feature = "use_gpu")] - pub unsafe fn set_samples(&self, index: u32, rs: RaySamples) { - if index >= self.capacity { - return; - } - let i = index as usize; - - let flags = if rs.have_subsurface { 1.0 } else { 0.0 }; - let dir = Float4 { - v: [rs.direct.u.0[0], rs.direct.u.0[1], rs.direct.uc, flags], - }; - - let ind = Float4 { - v: [ - rs.indirect.uc, - rs.indirect.rr, - rs.indirect.u.0[0], - rs.indirect.u.0[1], - ], - }; - - unsafe { - *self.rs_direct_packed.add(i) = dir; - *self.rs_indirect_packed.add(i) = ind; - } - - if rs.have_subsurface { - let ss = Float4 { - v: [ - rs.subsurface.uc, - rs.subsurface.u.0[0], - rs.subsurface.u.0[1], - 0.0, - ], - }; - unsafe { - *self.rs_subsurface_packed.add(i) = ss; - } - } - } -} diff --git a/shared/build.rs b/shared/build.rs new file mode 100644 index 0000000..104c9d6 --- /dev/null +++ b/shared/build.rs @@ -0,0 +1,5 @@ +fn main() { + // This allows "spirv" to be used in #[cfg(target_arch = "...")] + // without triggering a warning. + println!("cargo:rustc-check-cfg=cfg(target_arch, values(\"spirv\"))"); +} diff --git a/shared/src/bxdfs/layered.rs b/shared/src/bxdfs/layered.rs index d975b2e..48de890 100644 --- a/shared/src/bxdfs/layered.rs +++ b/shared/src/bxdfs/layered.rs @@ -19,7 +19,6 @@ use crate::spectra::{ N_SPECTRUM_SAMPLES, RGBColorSpace, RGBUnboundedSpectrum, SampledSpectrum, SampledWavelengths, StandardColorSpaces, }; -use crate::utils::DevicePtr; use crate::utils::hash::hash_buffer; use crate::utils::math::{ clamp, fast_exp, i0, lerp, log_i0, radians, safe_acos, safe_asin, safe_sqrt, sample_discrete, @@ -91,8 +90,8 @@ where thickness: Float, g: Float, albedo: SampledSpectrum, - max_depth: usize, - n_samples: usize, + max_depth: u32, + n_samples: u32, } impl LayeredBxDF @@ -106,8 +105,8 @@ where thickness: Float, albedo: SampledSpectrum, g: Float, - max_depth: usize, - n_samples: usize, + max_depth: u32, + n_samples: u32, ) -> Self { Self { top, diff --git a/shared/src/bxdfs/measured.rs b/shared/src/bxdfs/measured.rs index 8b733f9..51add6d 100644 --- a/shared/src/bxdfs/measured.rs +++ b/shared/src/bxdfs/measured.rs @@ -8,7 +8,7 @@ use crate::core::geometry::{ use crate::core::scattering::reflect; use crate::spectra::{SampledSpectrum, SampledWavelengths}; use crate::utils::math::square; -use crate::utils::ptr::{DevicePtr, Slice}; +use crate::utils::ptr::Ptr; use crate::utils::sampling::{PiecewiseLinear2D, cosine_hemisphere_pdf, sample_cosine_hemisphere}; use crate::{Float, INV_PI, PI, PI_OVER_2}; use core::any::Any; @@ -16,7 +16,7 @@ use core::any::Any; #[repr(C)] #[derive(Debug, Copy, Clone)] pub struct MeasuredBxDFData { - pub wavelengths: Slice, + pub wavelengths: Ptr, pub spectra: PiecewiseLinear2D<3>, pub ndf: PiecewiseLinear2D<0>, pub vndf: PiecewiseLinear2D<2>, @@ -28,7 +28,7 @@ pub struct MeasuredBxDFData { #[repr(C)] #[derive(Debug, Copy, Clone)] pub struct MeasuredBxDF { - pub brdf: DevicePtr, + pub brdf: Ptr, pub lambda: SampledWavelengths, } @@ -38,7 +38,7 @@ unsafe impl Sync for MeasuredBxDF {} impl MeasuredBxDF { pub fn new(brdf: &MeasuredBxDFData, lambda: &SampledWavelengths) -> Self { Self { - brdf: DevicePtr::from(brdf), + brdf: Ptr::from(brdf), lambda: *lambda, } } diff --git a/shared/src/cameras/mod.rs b/shared/src/cameras/mod.rs index 6076224..80822cf 100644 --- a/shared/src/cameras/mod.rs +++ b/shared/src/cameras/mod.rs @@ -6,4 +6,4 @@ mod spherical; pub use orthographic::OrthographicCamera; pub use perspective::PerspectiveCamera; pub use realistic::RealisticCamera; -pub use spherical::SphericalCamera; +pub use spherical::{SphericalCamera, Mapping}; diff --git a/shared/src/cameras/realistic.rs b/shared/src/cameras/realistic.rs index 9d41912..31957a6 100644 --- a/shared/src/cameras/realistic.rs +++ b/shared/src/cameras/realistic.rs @@ -11,6 +11,7 @@ use crate::core::pbrt::Float; use crate::core::sampler::CameraSample; use crate::core::scattering::refract; use crate::spectra::{SampledSpectrum, SampledWavelengths}; +use crate::utils::Ptr; use crate::utils::math::{lerp, quadratic, square}; #[repr(C)] @@ -37,8 +38,8 @@ pub struct RealisticCamera { base: CameraBase, focus_distance: Float, set_aperture_diameter: Float, - aperture_image: *const DeviceImage, - element_interfaces: *const LensElementInterface, + aperture_image: Ptr, + element_interfaces: Ptr, n_elements: usize, physical_extent: Bounds2f, exit_pupil_bounds: [Bounds2f; EXIT_PUPIL_SAMPLES], @@ -51,13 +52,13 @@ impl RealisticCamera { lens_params: &[Float], focus_distance: Float, set_aperture_diameter: Float, - aperture_image: Option, + aperture_image: Ptr, ) -> Self { let film_ptr = base.film; if film_ptr.is_null() { panic!("Camera must have a film"); } - let film = unsafe { &*film_ptr }; + let film = &*film_ptr; let aspect = film.full_resolution().x() as Float / film.full_resolution().y() as Float; let diagonal = film.diagonal(); @@ -90,7 +91,6 @@ impl RealisticCamera { element_interface.push(el_int); } - let n_samples = 64; let half_diag = film.diagonal() / 2.0; let mut exit_pupil_bounds = [Bounds2f::default(); EXIT_PUPIL_SAMPLES]; @@ -107,7 +107,7 @@ impl RealisticCamera { Self { base, focus_distance, - element_interfaces, + element_interfaces: Ptr::from(element_interfaces), n_elements, physical_extent, set_aperture_diameter, @@ -123,14 +123,16 @@ impl RealisticCamera { } pub fn compute_thick_lens_approximation(&self) -> ([Float; 2], [Float; 2]) { + use crate::utils::Ptr; + let x = 0.001 * self.get_film().diagonal(); let r_scene = Ray::new( Point3f::new(0., x, self.lens_front_z() + 1.), Vector3f::new(0., 0., -1.), None, - None, + &Ptr::null(), ); - let Some(r_film) = self.trace_lenses_from_film(r_scene) else { + let Some((_, r_film)) = self.trace_lenses_from_film(&r_scene) else { panic!( "Unable to trace ray from scene to film for thick lens approx. Is aperture very small?" ) @@ -140,14 +142,14 @@ impl RealisticCamera { Point3f::new(x, 0., self.lens_rear_z() - 1.), Vector3f::new(0., 0., 1.), None, - None, + &Ptr::null(), ); - let Some(r_scene) = self.trace_lenses_from_film(r_film) else { + let Some((_, r_scene)) = self.trace_lenses_from_film(&r_film) else { panic!( "Unable to trace ray from scene to film for thick lens approx. Is aperture very small?" ) }; - let (pz1, f_1) = Self::compute_cardinal_points(r_film, r_scene); + let (pz1, fz1) = Self::compute_cardinal_points(r_film, r_scene); ([pz0, pz1], [fz0, fz1]) } @@ -155,19 +157,23 @@ impl RealisticCamera { let (pz, fz) = self.compute_thick_lens_approximation(); let f = fz[0] - pz[0]; let z = -focus_distance; - let c = (pz[1] - z - pz[0]) * (pz[1] - z - 4 * f - pz[0]); - if c <= 0 { + let c = (pz[1] - z - pz[0]) * (pz[1] - z - 4. * f - pz[0]); + if c <= 0. { panic!( "Coefficient must be positive. It looks focusDistance {} is too short for a given lenses configuration", - focusDistance + focus_distance ); } let delta = (pz[1] - z + pz[0] - c.sqrt()) / 2.; - self.element_interface.last().thickness + delta + let last_interface = unsafe { self.element_interfaces.add(self.n_elements) }; + last_interface.thickness + delta } pub fn bound_exit_pupil(&self, film_x_0: Float, film_x_1: Float) -> Bounds2f { - Self::compute_exit_pupil_bounds(&self.element_interface, film_x_0, film_x_1) + let interface_array = unsafe { + core::slice::from_raw_parts(self.element_interfaces.as_raw(), self.n_elements as usize) + }; + Self::compute_exit_pupil_bounds(interface_array, film_x_0, film_x_1) } fn compute_exit_pupil_bounds( @@ -203,7 +209,10 @@ impl RealisticCamera { // Expand pupil bounds if ray makes it through the lens system if !pupil_bounds.contains(Point2f::new(p_rear.x(), p_rear.y())) - && trace_lenses_from_film(Ray::new(p_film, p_rear - p_film, None, None), None) + && trace_lenses_from_film( + Ray::new(p_film, p_rear - p_film, None, &Ptr::null()), + None, + ) { pupil_bounds = pupil_bounds.union_point(Point2f::new(p_rear.x(), p_rear.y())); } @@ -223,7 +232,7 @@ impl RealisticCamera { pub fn sample_exit_pupil(&self, p_film: Point2f, u_lens: Point2f) -> Option { // Find exit pupil bound for sample distance from film center - let film = self.film(); + let film = self.get_film(); let r_film = (square(p_film.x()) + square(p_film.y())).sqrt(); let mut r_index = (r_film / (film.diagonal() / 2.)) as usize * self.exit_pupil_bounds.len(); r_index = (self.exit_pupil_bounds.len() - 1).min(r_index); @@ -265,11 +274,11 @@ impl RealisticCamera { Point3f::new(r_camera.o.x(), r_camera.o.y(), -r_camera.o.z()), Vector3f::new(r_camera.d.x(), r_camera.d.y(), -r_camera.d.z()), Some(r_camera.time), - None, + &Ptr::null(), ); - for i in (0..self.element_interface.len() - 1).rev() { - let element: &LensElementInterface = &self.element_interface[i]; + for i in (0..self.n_elements - 1).rev() { + let element: &LensElementInterface = unsafe { &self.element_interfaces.add(i) }; // Update ray from film accounting for interaction with _element_ element_z -= element.thickness; @@ -308,8 +317,9 @@ impl RealisticCamera { // Update ray path for element interface interaction if !is_stop { let eta_i = element.eta; - let eta_t = if i > 0 && self.element_interface[i - 1].eta != 0. { - self.element_interface[i - 1].eta + let interface_i = unsafe { self.element_interfaces.add(i - 1) }; + let eta_t = if i > 0 && interface_i.eta != 0. { + interface_i.eta } else { 1. }; @@ -331,7 +341,7 @@ impl RealisticCamera { Point3f::new(r_lens.o.x(), r_lens.o.y(), -r_lens.o.z()), Vector3f::new(r_lens.d.x(), r_lens.d.y(), -r_lens.d.z()), Some(r_lens.time), - None, + &Ptr::null(), ); Some((weight, r_out)) @@ -368,19 +378,22 @@ impl RealisticCamera { } pub fn lens_rear_z(&self) -> Float { - self.element_interface.last().unwrap().thickness + let last_interface = unsafe { self.element_interfaces.add(self.n_elements - 1) }; + last_interface.thickness } pub fn lens_front_z(&self) -> Float { let mut z_sum = 0.; - for element in &self.element_interface { + for i in 0..self.n_elements { + let element = unsafe { self.element_interfaces.add(i) }; z_sum += element.thickness; } z_sum } pub fn rear_element_radius(&self) -> Float { - self.element_interface.last().unwrap().aperture_radius + let last_interface = unsafe { self.element_interfaces.add(self.n_elements - 1) }; + last_interface.aperture_radius } } @@ -406,7 +419,7 @@ impl CameraTrait for RealisticCamera { let eps = self.sample_exit_pupil(Point2f::new(p_film.x(), p_film.y()), sample.p_lens)?; let p_pupil = Point3f::new(0., 0., 0.); - let r_film = Ray::new(p_film, p_pupil - p_film, None, None); + let r_film = Ray::new(p_film, p_pupil - p_film, None, &Ptr::null()); let (weight, mut ray) = self.trace_lenses_from_film(&r_film)?; if weight == 0. { return None; diff --git a/shared/src/cameras/spherical.rs b/shared/src/cameras/spherical.rs index c46b34a..8c97a91 100644 --- a/shared/src/cameras/spherical.rs +++ b/shared/src/cameras/spherical.rs @@ -53,7 +53,7 @@ impl CameraTrait for SphericalCamera { Point3f::new(0., 0., 0.), dir, Some(self.sample_time(sample.time)), - self.base().medium.clone(), + &self.base().medium, ); Some(CameraRay { ray: self.render_from_camera(&ray, &mut None), diff --git a/shared/src/core/bsdf.rs b/shared/src/core/bsdf.rs index 470c0d8..4377a4b 100644 --- a/shared/src/core/bsdf.rs +++ b/shared/src/core/bsdf.rs @@ -2,17 +2,17 @@ use crate::Float; use crate::core::bxdf::{BSDFSample, BxDF, BxDFFlags, BxDFTrait, FArgs, TransportMode}; use crate::core::geometry::{Frame, Normal3f, Point2f, Vector3f, VectorLike}; use crate::spectra::SampledSpectrum; -use crate::utils::DevicePtr; +use crate::utils::Ptr; #[repr(C)] -#[derive(Copy, Debug, Default)] +#[derive(Copy, Clone, Debug, Default)] pub struct BSDF { - bxdf: DevicePtr, + bxdf: Ptr, shading_frame: Frame, } impl BSDF { - pub fn new(ns: Normal3f, dpdus: Vector3f, bxdf: DevicePtr) -> Self { + pub fn new(ns: Normal3f, dpdus: Vector3f, bxdf: Ptr) -> Self { Self { bxdf, shading_frame: Frame::new(dpdus.normalize(), Vector3f::from(ns)), @@ -66,7 +66,7 @@ impl BSDF { u2: Point2f, f_args: FArgs, ) -> Option { - let bxdf = self.bxdf.as_ref()?; + let bxdf = unsafe { self.bxdf.as_ref() }; let sampling_flags = BxDFFlags::from_bits_truncate(f_args.sample_flags.bits()); let wo = self.render_to_local(wo_render); @@ -119,7 +119,8 @@ impl BSDF { pub fn regularize(&mut self) { if !self.bxdf.is_null() { - unsafe { self.bxdf.as_mut().regularize() } + let mut bxdf = unsafe { *self.bxdf.as_raw() }; + bxdf.regularize(); } } } diff --git a/shared/src/core/bssrdf.rs b/shared/src/core/bssrdf.rs index fa755cc..616140d 100644 --- a/shared/src/core/bssrdf.rs +++ b/shared/src/core/bssrdf.rs @@ -1,4 +1,5 @@ -use crate::core::bxdf::{BSDF, NormalizedFresnelBxDF}; +use crate::bxdfs::NormalizedFresnelBxDF; +use crate::core::bsdf::BSDF; use crate::core::geometry::{Frame, Normal3f, Point2f, Point3f, Point3fi, Vector3f}; use crate::core::interaction::{InteractionBase, ShadingGeom, SurfaceInteraction}; use crate::core::shape::Shape; @@ -6,7 +7,6 @@ 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::{Ptr, ptr::Slice}; use crate::{Float, PI}; use enum_dispatch::enum_dispatch; use std::sync::Arc; @@ -105,30 +105,35 @@ pub struct BSSRDFTable { impl BSSRDFTable { pub fn get_rho(&self) -> &[Float] { - unsafe { core::slice::from_raw_parts(self.rho_samples.0, self.n_rho_samples as usize) } + unsafe { + core::slice::from_raw_parts(self.rho_samples.as_ref(), self.n_rho_samples as usize) + } } pub fn get_radius(&self) -> &[Float] { unsafe { - core::slice::from_raw_parts(self.radius_samples.0, self.n_radius_samples as usize) + core::slice::from_raw_parts( + self.radius_samples.as_ref(), + self.n_radius_samples as usize, + ) } } pub fn get_profile(&self) -> &[Float] { let n_profile = (self.n_rho_samples * self.n_radius_samples) as usize; - unsafe { core::slice::from_raw_parts(self.profile.0, n_profile) } + unsafe { core::slice::from_raw_parts(self.profile.as_ref(), n_profile) } } pub fn get_cdf(&self) -> &[Float] { let n_profile = (self.n_rho_samples * self.n_radius_samples) as usize; - unsafe { core::slice::from_raw_parts(self.profile_cdf.0, n_profile) } + unsafe { core::slice::from_raw_parts(self.profile_cdf.as_ref(), n_profile) } } pub fn eval_profile(&self, rho_index: u32, radius_index: u32) -> Float { debug_assert!(rho_index < self.n_rho_samples); debug_assert!(radius_index < self.n_radius_samples); let idx = (rho_index * self.n_radius_samples + radius_index) as usize; - unsafe { *self.profile.0.add(idx) } + unsafe { *self.profile.add(idx) } } } diff --git a/shared/src/core/camera.rs b/shared/src/core/camera.rs index 4c1b019..e1328ed 100644 --- a/shared/src/core/camera.rs +++ b/shared/src/core/camera.rs @@ -10,7 +10,7 @@ use crate::core::pbrt::Float; use crate::core::sampler::CameraSample; use crate::spectra::{SampledSpectrum, SampledWavelengths}; use crate::utils::math::lerp; -use crate::utils::ptr::DevicePtr; +use crate::utils::ptr::Ptr; use crate::utils::transform::{AnimatedTransform, Transform}; use enum_dispatch::enum_dispatch; @@ -113,8 +113,8 @@ pub struct CameraBase { pub min_pos_differential_y: Vector3f, pub min_dir_differential_x: Vector3f, pub min_dir_differential_y: Vector3f, - pub film: DevicePtr, - pub medium: DevicePtr, + pub film: Ptr, + pub medium: Ptr, } #[enum_dispatch(CameraTrait)] @@ -163,49 +163,44 @@ pub trait CameraTrait { sample: CameraSample, lambda: &SampledWavelengths, ) -> Option { - match self { - Camera::Orthographic(c) => c.generate_ray_differential(sample, lambda), - _ => { - let mut central_cam_ray = self.generate_ray(sample, lambda)?; - let mut rd = RayDifferential::default(); - let mut rx_found = false; - let mut ry_found = false; + let mut central_cam_ray = self.generate_ray(sample, lambda)?; + let mut rd = RayDifferential::default(); + let mut rx_found = false; + let mut ry_found = false; - for eps in [0.05, -0.05] { - let mut s_shift = sample; - s_shift.p_film[0] += eps; + for eps in [0.05, -0.05] { + let mut s_shift = sample; + s_shift.p_film[0] += eps; - if let Some(rx_cam_ray) = self.generate_ray(s_shift, lambda) { - rd.rx_origin = central_cam_ray.ray.o - + (rx_cam_ray.ray.o - central_cam_ray.ray.o) / eps; - rd.rx_direction = central_cam_ray.ray.d - + (rx_cam_ray.ray.d - central_cam_ray.ray.d) / eps; - rx_found = true; - break; - } - } - - for eps in [0.05, -0.05] { - let mut s_shift = sample; - s_shift.p_film[1] += eps; - - if let Some(ry_cam_ray) = self.generate_ray(s_shift, lambda) { - rd.ry_origin = central_cam_ray.ray.o - + (ry_cam_ray.ray.o - central_cam_ray.ray.o) / eps; - rd.ry_direction = central_cam_ray.ray.d - + (ry_cam_ray.ray.d - central_cam_ray.ray.d) / eps; - ry_found = true; - break; - } - } - - if rx_found && ry_found { - central_cam_ray.ray.differential = rd; - } - - Some(central_cam_ray) + if let Some(rx_cam_ray) = self.generate_ray(s_shift, lambda) { + rd.rx_origin = + central_cam_ray.ray.o + (rx_cam_ray.ray.o - central_cam_ray.ray.o) / eps; + rd.rx_direction = + central_cam_ray.ray.d + (rx_cam_ray.ray.d - central_cam_ray.ray.d) / eps; + rx_found = true; + break; } } + + for eps in [0.05, -0.05] { + let mut s_shift = sample; + s_shift.p_film[1] += eps; + + if let Some(ry_cam_ray) = self.generate_ray(s_shift, lambda) { + rd.ry_origin = + central_cam_ray.ray.o + (ry_cam_ray.ray.o - central_cam_ray.ray.o) / eps; + rd.ry_direction = + central_cam_ray.ray.d + (ry_cam_ray.ray.d - central_cam_ray.ray.d) / eps; + ry_found = true; + break; + } + } + + if rx_found && ry_found { + central_cam_ray.ray.differential = rd; + } + + Some(central_cam_ray) } fn approximate_dp_dxy( @@ -232,13 +227,13 @@ pub trait CameraTrait { Point3f::new(0., 0., 0.) + self.base().min_pos_differential_x, Vector3f::new(0., 0., 1.) + self.base().min_dir_differential_x, None, - &DevicePtr::default(), + &Ptr::default(), ); let y_ray = Ray::new( Point3f::new(0., 0., 0.) + self.base().min_pos_differential_y, Vector3f::new(0., 0., 1.) + self.base().min_dir_differential_y, None, - &DevicePtr::default(), + &Ptr::default(), ); let n_down = Vector3f::from(n_down_z); let tx = -(n_down.dot(y_ray.o.into())) / n_down.dot(x_ray.d); diff --git a/shared/src/core/color.rs b/shared/src/core/color.rs index 6e81558..eeb3776 100644 --- a/shared/src/core/color.rs +++ b/shared/src/core/color.rs @@ -27,8 +27,8 @@ 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) + fn from(triplet: [Float; 3]) -> Self { + XYZ::new(triplet[0], triplet[1], triplet[2]) } } @@ -336,6 +336,30 @@ impl Index for RGB { } } +impl Index for RGB { + type Output = Float; + fn index(&self, index: i32) -> &Self::Output { + debug_assert!(index < 3); + match index { + 0 => &self.r, + 1 => &self.g, + _ => &self.b, + } + } +} + +impl Index for RGB { + type Output = Float; + fn index(&self, index: usize) -> &Self::Output { + debug_assert!(index < 3); + match index { + 0 => &self.r, + 1 => &self.g, + _ => &self.b, + } + } +} + impl IndexMut for RGB { fn index_mut(&mut self, index: u32) -> &mut Self::Output { debug_assert!(index < 3); @@ -347,6 +371,28 @@ impl IndexMut for RGB { } } +impl IndexMut for RGB { + fn index_mut(&mut self, index: i32) -> &mut Self::Output { + debug_assert!(index < 3); + match index { + 0 => &mut self.r, + 1 => &mut self.g, + _ => &mut self.b, + } + } +} + +impl IndexMut for RGB { + fn index_mut(&mut self, index: usize) -> &mut Self::Output { + debug_assert!(index < 3); + match index { + 0 => &mut self.r, + 1 => &mut self.g, + _ => &mut self.b, + } + } +} + impl Neg for RGB { type Output = Self; fn neg(self) -> Self { @@ -667,7 +713,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 u32]; + vout[i] = SRGB_TO_LINEAR_LUT[v as usize]; } } @@ -992,6 +1038,18 @@ impl Add for Coeffs { } } +impl Sub for Coeffs { + type Output = Self; + #[inline(always)] + fn sub(self, rhs: Self) -> Self { + Self { + c0: self.c0 - rhs.c0, + c1: self.c1 - rhs.c1, + c2: self.c2 - rhs.c2, + } + } +} + impl Mul for Coeffs { type Output = Self; #[inline(always)] @@ -1025,7 +1083,7 @@ impl RGBToSpectrumTable { let m = rgb.max_component_value(); let min_val = rgb.min_component_value(); if m - min_val < 1e-4 { - let x = clamp(rgb[0], 1e-4, 0.9999); + let x: Float = clamp(rgb[0], 1e-4, 0.9999); let c2 = (0.5 - x) / (x * (1.0 - x)).sqrt(); return RGBSigmoidPolynomial::new(0.0, 0.0, c2); } diff --git a/shared/src/core/film.rs b/shared/src/core/film.rs index 790647b..f878065 100644 --- a/shared/src/core/film.rs +++ b/shared/src/core/film.rs @@ -13,16 +13,15 @@ use crate::spectra::{ ConstantSpectrum, DenselySampledSpectrum, LAMBDA_MAX, LAMBDA_MIN, N_SPECTRUM_SAMPLES, PiecewiseLinearSpectrum, RGBColorSpace, SampledSpectrum, SampledWavelengths, colorspace, }; -use crate::utils::AtomicFloat; use crate::utils::containers::Array2D; use crate::utils::math::linear_least_squares; use crate::utils::math::{SquareMatrix, wrap_equal_area_square}; -use crate::utils::ptr::DevicePtr; use crate::utils::sampling::VarianceEstimator; use crate::utils::transform::AnimatedTransform; +use crate::utils::{AtomicFloat, Ptr}; #[repr(C)] -#[derive(Clone, Copy, Debug)] +#[derive(Debug)] pub struct RGBFilm { pub base: FilmBase, pub max_component_value: Float, @@ -33,7 +32,7 @@ pub struct RGBFilm { } #[repr(C)] -#[derive(Clone, Debug)] +#[derive(Debug)] pub struct RGBPixel { rgb_sum: [AtomicFloat; 3], weight_sum: AtomicFloat, @@ -108,7 +107,7 @@ impl RGBFilm { _vi: Option<&VisibleSurface>, weight: Float, ) { - let sensor = unsafe { self.get_sensor() }; + let sensor = self.get_sensor(); let mut rgb = sensor.to_sensor_rgb(l, lambda); let m = rgb.into_iter().copied().fold(f32::NEG_INFINITY, f32::max); if m > self.max_component_value { @@ -117,13 +116,13 @@ impl RGBFilm { let pixel = &self.pixels[p_film]; for c in 0..3 { - pixel.rgb_sum[c].add((weight * rgb[c]) as f64); + pixel.rgb_sum[c].add((weight * rgb[c as u32]) as f32); } - pixel.weight_sum.add(weight as f64); + pixel.weight_sum.add(weight as f32); } pub fn add_splat(&mut self, p: Point2f, l: SampledSpectrum, lambda: &SampledWavelengths) { - let sensor = unsafe { self.get_sensor() }; + let sensor = self.get_sensor(); let mut rgb = sensor.to_sensor_rgb(l, lambda); let m = rgb.into_iter().copied().fold(f32::NEG_INFINITY, f32::max); if m > self.max_component_value { @@ -148,32 +147,32 @@ impl RGBFilm { if wt != 0. { let pixel = &self.pixels[*pi]; for i in 0..3 { - pixel.rgb_splat[i].add((wt * rgb[i]) as f64); + pixel.rgb_splat[i].add((wt * rgb[i as u32]) as f32); } } } } pub fn get_pixel_rgb(&self, p: Point2i, splat_scale: Option) -> RGB { - let pixel = unsafe { &self.pixels.get(p) }; + let pixel = &self.pixels.get(p); let mut rgb = RGB::new( - pixel.rgb_sum[0].load() as Float, - pixel.rgb_sum[1].load() as Float, - pixel.rgb_sum[2].load() as Float, + pixel.rgb_sum[0].get() as Float, + pixel.rgb_sum[1].get() as Float, + pixel.rgb_sum[2].get() as Float, ); - let weight_sum = pixel.weight_sum.load(); + let weight_sum = pixel.weight_sum.get(); if weight_sum != 0. { rgb /= weight_sum as Float } if let Some(splat) = splat_scale { for c in 0..3 { - let splat_val = pixel.rgb_splat[c].load(); + let splat_val = pixel.rgb_splat[c].get(); rgb[c] += splat * splat_val as Float / self.filter_integral; } } else { for c in 0..3 { - let splat_val = pixel.rgb_splat[c].load(); + let splat_val = pixel.rgb_splat[c].get(); rgb[c] += splat_val as Float / self.filter_integral; } } @@ -210,7 +209,7 @@ struct GBufferPixel { } #[repr(C)] -#[derive(Debug, Copy)] +#[derive(Debug)] #[cfg_attr(target_os = "cuda", derive(Copy, Clone))] pub struct GBufferFilm { pub base: FilmBase, @@ -224,40 +223,6 @@ pub struct GBufferFilm { output_rgbf_from_sensor_rgb: SquareMatrix, } -#[cfg(not(target_os = "cuda"))] -impl GBufferFilm { - pub fn new( - base: &FilmBase, - output_from_render: &AnimatedTransform, - apply_inverse: bool, - colorspace: &RGBColorSpace, - max_component_value: Float, - write_fp16: bool, - ) -> Self { - assert!(!base.pixel_bounds.is_empty()); - let sensor_ptr = base.sensor; - if sensor_ptr.is_null() { - panic!("Film must have a sensor"); - } - let sensor = unsafe { &*sensor_ptr }; - let output_rgbf_from_sensor_rgb = colorspace.rgb_from_xyz * sensor.xyz_from_sensor_rgb; - let filter_integral = base.filter.integral(); - let pixels = Array2D::new(base.pixel_bounds); - - Self { - base: base.clone(), - output_from_render: output_from_render.clone(), - apply_inverse, - pixels, - colorspace: colorspace.clone(), - max_component_value, - write_fp16, - filter_integral, - output_rgbf_from_sensor_rgb, - } - } -} - impl GBufferFilm { pub fn base(&self) -> &FilmBase { &self.base @@ -279,8 +244,19 @@ impl GBufferFilm { unsafe { &*self.base.sensor } } + pub fn add_sample( + &mut self, + _p_film: Point2i, + _l: SampledSpectrum, + _lambda: &SampledWavelengths, + _visible_surface: Option<&VisibleSurface>, + _weight: Float, + ) { + todo!() + } + pub fn add_splat(&mut self, p: Point2f, l: SampledSpectrum, lambda: &SampledWavelengths) { - let sensor = unsafe { self.get_sensor() }; + let sensor = self.get_sensor(); let mut rgb = sensor.to_sensor_rgb(l, lambda); let m = rgb.into_iter().copied().fold(f32::NEG_INFINITY, f32::max); if m > self.max_component_value { @@ -305,7 +281,7 @@ impl GBufferFilm { if wt != 0. { let pixel = &self.pixels[*pi]; for i in 0..3 { - pixel.rgb_splat[i].add((wt * rgb[i]) as f64); + pixel.rgb_splat[i].add((wt * rgb[i]) as f32); } } } @@ -318,25 +294,25 @@ impl GBufferFilm { } pub fn get_pixel_rgb(&self, p: Point2i, splat_scale: Option) -> RGB { - let pixel = unsafe { &self.pixels.get(p) }; + let pixel = &self.pixels.get(p); let mut rgb = RGB::new( - pixel.rgb_sum[0].load() as Float, - pixel.rgb_sum[1].load() as Float, - pixel.rgb_sum[2].load() as Float, + pixel.rgb_sum[0].get() as Float, + pixel.rgb_sum[1].get() as Float, + pixel.rgb_sum[2].get() as Float, ); - let weight_sum = pixel.weight_sum.load(); + let weight_sum = pixel.weight_sum.get(); if weight_sum != 0. { rgb /= weight_sum as Float } if let Some(splat) = splat_scale { for c in 0..3 { - let splat_val = pixel.rgb_splat[c].load(); + let splat_val = pixel.rgb_splat[c].get(); rgb[c] += splat * splat_val as Float / self.filter_integral; } } else { for c in 0..3 { - let splat_val = pixel.rgb_splat[c].load(); + let splat_val = pixel.rgb_splat[c].get(); rgb[c] += splat_val as Float / self.filter_integral; } } @@ -359,7 +335,7 @@ pub struct SpectralPixel { } #[repr(C)] -#[derive(Debug, Default)] +#[derive(Debug)] #[cfg_attr(target_os = "cuda", derive(Copy, Clone))] pub struct SpectralFilm { pub base: FilmBase, @@ -389,6 +365,21 @@ impl SpectralFilm { fn uses_visible_surface(&self) -> bool { true } + + pub fn add_sample( + &mut self, + _p_film: Point2i, + _l: SampledSpectrum, + _lambda: &SampledWavelengths, + _visible_surface: Option<&VisibleSurface>, + _weight: Float, + ) { + todo!() + } + + pub fn add_splat(&mut self, _p: Point2f, _v: SampledSpectrum, _lambda: &SampledWavelengths) { + todo!() + } } #[repr(C)] @@ -402,102 +393,6 @@ pub struct PixelSensor { } impl PixelSensor { - const N_SWATCH_REFLECTANCES: usize = 24; - #[cfg(not(target_os = "cuda"))] - pub fn new( - r: Spectrum, - g: Spectrum, - b: Spectrum, - output_colorspace: RGBColorSpace, - sensor_illum: &Spectrum, - imaging_ratio: Float, - spectra: *const StandardSpectra, - swatches: &[Spectrum; 24], - ) -> Self { - // As seen in usages of this constructos, sensor_illum can be null - // Going with the colorspace's own illuminant, but this might not be the right choice - // TODO: Test this - let illum: &Spectrum = match sensor_illum { - Some(arc_illum) => &**arc_illum, - None => &output_colorspace.illuminant, - }; - - let r_bar = DenselySampledSpectrum::from_spectrum(&r); - let g_bar = DenselySampledSpectrum::from_spectrum(&g); - let b_bar = DenselySampledSpectrum::from_spectrum(&b); - let mut rgb_camera = [[0.; 3]; Self::N_SWATCH_REFLECTANCES]; - - let swatches = Self::get_swatches(); - - for i in 0..Self::N_SWATCH_REFLECTANCES { - let rgb = Self::project_reflectance::( - &swatches[i], - illum, - &Spectrum::DenselySampled(r_bar.clone()), - &Spectrum::DenselySampled(g_bar.clone()), - &Spectrum::DenselySampled(b_bar.clone()), - ); - for c in 0..3 { - rgb_camera[i][c] = rgb[c]; - } - } - - let mut xyz_output = [[0.; 3]; Self::N_SWATCH_REFLECTANCES]; - let sensor_white_g = illum.inner_product(&Spectrum::DenselySampled(g_bar.clone())); - let sensor_white_y = illum.inner_product(spectra.y); - for i in 0..Self::N_SWATCH_REFLECTANCES { - let s = swatches[i].clone(); - let xyz = Self::project_reflectance::( - &s, - &output_colorspace.illuminant, - spectra.x, - spectra.y, - spectra.z, - ) * (sensor_white_y / sensor_white_g); - for c in 0..3 { - xyz_output[i][c] = xyz[c]; - } - } - - let xyz_from_sensor_rgb = linear_least_squares(rgb_camera, xyz_output)?; - - Ok(Self { - xyz_from_sensor_rgb, - r_bar, - g_bar, - b_bar, - imaging_ratio, - }) - } - - pub fn new_with_white_balance( - output_colorspace: &RGBColorSpace, - sensor_illum: DevicePtr, - imaging_ratio: Float, - spectra: *const StandardSpectra, - ) -> Self { - let r_bar = DenselySampledSpectrum::from_spectrum(spectra.x); - let g_bar = DenselySampledSpectrum::from_spectrum(spectra.y); - let b_bar = DenselySampledSpectrum::from_spectrum(spectra.z); - let xyz_from_sensor_rgb: SquareMatrix; - - if let Some(illum) = sensor_illum { - let source_white = illum.to_xyz(spectra).xy(); - let target_white = output_colorspace.w; - xyz_from_sensor_rgb = white_balance(source_white, target_white); - } else { - xyz_from_sensor_rgb = SquareMatrix::::default(); - } - - Self { - xyz_from_sensor_rgb, - r_bar, - g_bar, - b_bar, - imaging_ratio, - } - } - pub fn project_reflectance( refl: &Spectrum, illum: &Spectrum, @@ -529,7 +424,7 @@ impl PixelSensor { result[2] *= inv_g; } - T::from((result[0], result[1], result[2])) + T::from([result[0], result[1], result[2]]) } } @@ -609,7 +504,7 @@ impl Film { } pub fn add_sample( - &self, + &mut self, p_film: Point2i, l: SampledSpectrum, lambda: &SampledWavelengths, diff --git a/shared/src/core/filter.rs b/shared/src/core/filter.rs index 4d97c14..854a4f9 100644 --- a/shared/src/core/filter.rs +++ b/shared/src/core/filter.rs @@ -20,33 +20,6 @@ pub struct FilterSampler { } impl FilterSampler { - #[cfg(not(target_os = "cuda"))] - pub fn new(radius: Vector2f, func: F) -> Self - where - F: Fn(Point2f) -> Float, - { - let domain = Bounds2f::from_points( - Point2f::new(-radius.x(), -radius.y()), - Point2f::new(radius.x(), radius.y()), - ); - - let nx = (32.0 * radius.x()) as usize; - let ny = (32.0 * radius.y()) as usize; - - let mut f = Array2D::new_with_dims(nx, ny); - for y in 0..f.y_size() { - for x in 0..f.x_size() { - let p = domain.lerp(Point2f::new( - (x as Float + 0.5) / f.x_size() as Float, - (y as Float + 0.5) / f.y_size() as Float, - )); - f[(x as i32, y as i32)] = func(p); - } - } - let distrib = DevicePiecewiseConstant2D::new_with_bounds(&f, domain); - Self { domain, f, distrib } - } - pub fn sample(&self, u: Point2f) -> FilterSample { let (p, pdf, pi) = self.distrib.sample(u); @@ -54,7 +27,8 @@ impl FilterSampler { return FilterSample { p, weight: 0.0 }; } - let weight = *self.f.get_linear(pi.x() as u32 + self.f.x_size()) / pdf; + let idx = pi.x() as u32 + self.f.x_size(); + let weight = *self.f.get_linear(idx as usize) / pdf; FilterSample { p, weight } } } diff --git a/shared/src/core/geometry/ray.rs b/shared/src/core/geometry/ray.rs index d51b97e..1daf9c0 100644 --- a/shared/src/core/geometry/ray.rs +++ b/shared/src/core/geometry/ray.rs @@ -2,7 +2,7 @@ use super::{Normal3f, Point3f, Point3fi, Vector3f, VectorLike}; use crate::core::medium::Medium; use crate::core::pbrt::Float; use crate::utils::math::{next_float_down, next_float_up}; -use crate::utils::ptr::DevicePtr; +use crate::utils::ptr::Ptr; #[repr(C)] #[derive(Clone, Copy, Debug)] @@ -10,7 +10,7 @@ pub struct Ray { pub o: Point3f, pub d: Vector3f, pub time: Float, - pub medium: DevicePtr, + pub medium: Ptr, // We do this instead of creating a trait for Rayable or some gnarly thing like that pub has_differentials: bool, pub differential: RayDifferential, @@ -21,7 +21,7 @@ impl Default for Ray { Self { o: Point3f::new(0.0, 0.0, 0.0), d: Vector3f::new(0.0, 0.0, 0.0), - medium: DevicePtr::null(), + medium: Ptr::null(), time: 0.0, has_differentials: false, differential: RayDifferential::default(), @@ -35,7 +35,7 @@ impl Ray { o, d, time: time.unwrap_or_else(|| Self::default().time), - medium: DevicePtr::from(medium), + medium: Ptr::from(medium), ..Self::default() } } @@ -71,7 +71,7 @@ impl Ray { o: origin, d, time, - medium: DevicePtr::null(), + medium: Ptr::null(), has_differentials: false, differential: RayDifferential::default(), } @@ -99,7 +99,7 @@ impl Ray { o: pf, d, time, - medium: DevicePtr::null(), + medium: Ptr::null(), has_differentials: false, differential: RayDifferential::default(), } diff --git a/shared/src/core/image.rs b/shared/src/core/image.rs index 8185ad7..49954cb 100644 --- a/shared/src/core/image.rs +++ b/shared/src/core/image.rs @@ -35,6 +35,16 @@ pub enum PixelFormat { F32, } +impl std::fmt::Display for PixelFormat { + fn fmt(&self, f: &mut std::fmt::Formatter<'_>) -> std::fmt::Result { + match self { + PixelFormat::U8 => write!(f, "U256"), + PixelFormat::F16 => write!(f, "Half"), + PixelFormat::F32 => write!(f, "Float"), + } + } +} + impl PixelFormat { pub fn is_8bit(&self) -> bool { matches!(self, PixelFormat::U8) diff --git a/shared/src/core/interaction.rs b/shared/src/core/interaction.rs index 0d5cc25..84007f0 100644 --- a/shared/src/core/interaction.rs +++ b/shared/src/core/interaction.rs @@ -1,6 +1,8 @@ use crate::Float; +use crate::bxdfs::DiffuseBxDF; +use crate::core::bsdf::BSDF; use crate::core::bssrdf::BSSRDF; -use crate::core::bxdf::{BSDF, BxDF, BxDFFlags, DiffuseBxDF}; +use crate::core::bxdf::{BxDF, BxDFFlags}; use crate::core::camera::{Camera, CameraTrait}; use crate::core::geometry::{ Normal3f, Point2f, Point3f, Point3fi, Ray, RayDifferential, Vector3f, VectorLike, @@ -250,8 +252,8 @@ impl SurfaceInteraction { } pub fn compute_differentials(&mut self, r: &Ray, camera: &Camera, samples_per_pixel: i32) { - let computed = if !r.differential.is_null() { - let diff = unsafe { &*r.differential }; + let computed = if !r.has_differentials { + let diff = r.differential; let dot_rx = self.common.n.dot(diff.rx_direction.into()); let dot_ry = self.common.n.dot(diff.ry_direction.into()); @@ -338,8 +340,8 @@ impl SurfaceInteraction { let new_ray = Ray::spawn(&self.pi(), &self.n(), ray.time, ray.d); ray.o = new_ray.o; // Skipping other variables, since they should not change when passing through surface - if !ray.differential.is_null() { - let diff = unsafe { &mut *ray.differential }; + if !ray.has_differentials { + let mut diff = ray.differential; diff.rx_origin += diff.rx_direction * t; diff.ry_origin += diff.ry_direction * t; } @@ -357,12 +359,12 @@ impl SurfaceInteraction { let material = { let root_mat = self.material; - let mut active_mat: &Material = *root_mat; + let mut active_mat: &Material = &*root_mat; let tex_eval = UniversalTextureEvaluator; while let Material::Mix(mix) = active_mat { // We need a context to evaluate the 'amount' texture let ctx = MaterialEvalContext::from(&*self); - active_mat = mix.choose_material(&tex_eval, &ctx); + active_mat = mix.choose_material(&tex_eval, &ctx)?; } active_mat.clone() }; @@ -370,8 +372,8 @@ impl SurfaceInteraction { let ctx = MaterialEvalContext::from(&*self); let tex_eval = UniversalTextureEvaluator; let displacement = material.get_displacement(); - let normal_map = material.get_normal_map(); - if displacement.is_some() || normal_map.is_some() { + let normal_map = Ptr::from(material.get_normal_map().unwrap()); + if !displacement.is_null() || !normal_map.is_null() { // This calls the function defined above self.compute_bump_geom(&tex_eval, displacement, normal_map); } @@ -380,7 +382,7 @@ impl SurfaceInteraction { if get_options().force_diffuse { let r = bsdf.rho_wo(self.common.wo, &[sampler.get1d()], &[sampler.get2d()]); let diff_bxdf = BxDF::Diffuse(DiffuseBxDF::new(r)); - bsdf = BSDF::new(self.shading.n, self.shading.dpdu, Some(diff_bxdf)); + bsdf = BSDF::new(self.shading.n, self.shading.dpdu, Ptr::from(&diff_bxdf)); } Some(bsdf) } @@ -393,13 +395,12 @@ impl SurfaceInteraction { _camera: &Camera, ) -> Option { let material = { - let root_mat = self.material.as_deref()?; - let mut active_mat: &Material = root_mat; + let mut active_mat = unsafe { self.material.as_ref() }; let tex_eval = UniversalTextureEvaluator; while let Material::Mix(mix) = active_mat { // We need a context to evaluate the 'amount' texture let ctx = MaterialEvalContext::from(self); - active_mat = mix.choose_material(&tex_eval, &ctx); + active_mat = mix.choose_material(&tex_eval, &ctx)?; } active_mat.clone() }; @@ -418,8 +419,9 @@ impl SurfaceInteraction { let ctx = NormalBumpEvalContext::from(&*self); let (dpdu, dpdv) = if !displacement.is_null() { bump_map(tex_eval, &displacement, &ctx) - } else if let Some(map) = normal_image { - normal_map(map.as_ref(), &ctx) + } else if !normal_image.is_null() { + let map = unsafe { normal_image.as_ref() }; + normal_map(map, &ctx) } else { (self.shading.dpdu, self.shading.dpdv) }; @@ -441,7 +443,8 @@ impl SurfaceInteraction { ) -> Ray { let mut rd = self.spawn_ray(wi); - if let Some(diff_i) = &ray_i.differential { + if ray_i.has_differentials { + let diff_i = ray_i.differential; let mut n = self.shading.n; let mut dndx = self.shading.dndu * self.dudx + self.shading.dndv * self.dvdx; diff --git a/shared/src/core/light.rs b/shared/src/core/light.rs index c5bab66..a2a1762 100644 --- a/shared/src/core/light.rs +++ b/shared/src/core/light.rs @@ -17,7 +17,6 @@ use crate::spectra::{ }; use crate::utils::Transform; use crate::utils::math::{equal_area_sphere_to_square, radians, safe_sqrt, smooth_step, square}; -use crate::utils::ptr::{DevicePtr, Ptr}; use crate::utils::sampling::DevicePiecewiseConstant2D; use crate::{Float, PI}; use bitflags::bitflags; diff --git a/shared/src/core/material.rs b/shared/src/core/material.rs index 93756e5..e59263e 100644 --- a/shared/src/core/material.rs +++ b/shared/src/core/material.rs @@ -19,7 +19,7 @@ use crate::core::texture::{ }; use crate::materials::*; use crate::spectra::{SampledSpectrum, SampledWavelengths}; -use crate::utils::DevicePtr; +use crate::utils::Ptr; use crate::utils::hash::hash_float; use crate::utils::math::clamp; @@ -125,7 +125,7 @@ pub fn bump_map( displacement: &GPUFloatTexture, ctx: &NormalBumpEvalContext, ) -> (Vector3f, Vector3f) { - debug_assert!(tex_eval.can_evaluate(&[DevicePtr::from(displacement)], &[])); + debug_assert!(tex_eval.can_evaluate(&[Ptr::from(displacement)], &[])); let mut du = 0.5 * (ctx.dudx.abs() + ctx.dudy.abs()); if du == 0.0 { du = 0.0005; @@ -174,7 +174,7 @@ pub trait MaterialTrait { fn can_evaluate_textures(&self, tex_eval: &dyn TextureEvaluator) -> bool; fn get_normal_map(&self) -> Option<&DeviceImage>; - fn get_displacement(&self) -> DevicePtr; + fn get_displacement(&self) -> Ptr; fn has_subsurface_scattering(&self) -> bool; } diff --git a/shared/src/core/medium.rs b/shared/src/core/medium.rs index f318a46..ee704ff 100644 --- a/shared/src/core/medium.rs +++ b/shared/src/core/medium.rs @@ -12,7 +12,7 @@ use crate::spectra::{ }; use crate::utils::containers::SampledGrid; use crate::utils::math::{clamp, square}; -use crate::utils::ptr::DevicePtr; +use crate::utils::ptr::Ptr; use crate::utils::rng::Rng; use crate::utils::transform::Transform; @@ -94,22 +94,26 @@ impl PhaseFunctionTrait for HGPhaseFunction { pub struct MajorantGrid { pub bounds: Bounds3f, pub res: Point3i, - pub voxels: *const Float, + pub voxels: *mut Float, + pub n_voxels: u32, } unsafe impl Send for MajorantGrid {} unsafe impl Sync for MajorantGrid {} impl MajorantGrid { - #[cfg(not(target_os = "cuda"))] - pub fn new(bounds: Bounds3f, res: Point3i) -> Self { - Self { - bounds, - res, - voxels: Vec::with_capacity((res.x() * res.y() * res.z()) as usize), - } - } - + // #[cfg(not(target_os = "cuda"))] + // pub fn new(bounds: Bounds3f, res: Point3i) -> Self { + // let n_voxels = (res.x() * res.y() * res.z()) as usize; + // let voxels = Vec::with_capacity(n_voxels); + // Self { + // bounds, + // res, + // voxels: voxels.as_ptr(), + // n_voxels: n_voxels as u32, + // } + // } + // #[inline(always)] fn is_valid(&self) -> bool { !self.voxels.is_null() @@ -123,7 +127,7 @@ impl MajorantGrid { let idx = z * self.res.x() * self.res.y() + y * self.res.x() + x; - if idx >= 0 && (idx as usize) < self.voxels.len() { + if idx >= 0 && (idx as u32) < self.n_voxels { unsafe { *self.voxels.add(idx as usize) } } else { 0.0 @@ -131,7 +135,7 @@ impl MajorantGrid { } #[inline(always)] - pub fn set(&self, x: i32, y: i32, z: i32, v: Float) { + pub fn set(&mut self, x: i32, y: i32, z: i32, v: Float) { if !self.is_valid() { return; } @@ -264,7 +268,7 @@ impl DDAMajorantIterator { let p_grid_start = grid.bounds.offset(&ray.at(t_min)); let grid_intersect = Vector3f::from(p_grid_start); - let res = [grid.res.x, grid.res.y, grid.res.z]; + let res = [grid.res.x(), grid.res.y(), grid.res.z()]; for axis in 0..3 { iter.voxel[axis] = clamp( @@ -449,36 +453,10 @@ pub enum Medium { #[repr(C)] #[derive(Debug, Clone, Copy)] pub struct HomogeneousMedium { - sigma_a_spec: DenselySampledSpectrum, - sigma_s_spec: DenselySampledSpectrum, - le_spec: DenselySampledSpectrum, - phase: HGPhaseFunction, -} - -impl HomogeneousMedium { - pub fn new( - sigma_a: Spectrum, - sigma_s: Spectrum, - sigma_scale: Float, - le: Spectrum, - le_scale: Float, - g: Float, - ) -> Self { - let mut sigma_a_spec = DenselySampledSpectrum::from_spectrum(&sigma_a); - let mut sigma_s_spec = DenselySampledSpectrum::from_spectrum(&sigma_s); - let mut le_spec = DenselySampledSpectrum::from_spectrum(&le); - - sigma_a_spec.scale(sigma_scale); - sigma_s_spec.scale(sigma_scale); - le_spec.scale(le_scale); - - Self { - sigma_a_spec, - sigma_s_spec, - le_spec, - phase: HGPhaseFunction::new(g), - } - } + pub sigma_a_spec: DenselySampledSpectrum, + pub sigma_s_spec: DenselySampledSpectrum, + pub le_spec: DenselySampledSpectrum, + pub phase: HGPhaseFunction, } impl MediumTrait for HomogeneousMedium { @@ -517,70 +495,17 @@ impl MediumTrait for HomogeneousMedium { #[repr(C)] #[derive(Debug, Clone, Copy)] pub struct GridMedium { - bounds: Bounds3f, - render_from_medium: Transform, - sigma_a_spec: DenselySampledSpectrum, - sigma_s_spec: DenselySampledSpectrum, - density_grid: SampledGrid, - phase: HGPhaseFunction, - temperature_grid: SampledGrid, - le_spec: DenselySampledSpectrum, - le_scale: SampledGrid, - is_emissive: bool, - majorant_grid: MajorantGrid, -} - -impl GridMedium { - #[allow(clippy::too_many_arguments)] - #[cfg(not(target_os = "cuda"))] - pub fn new( - bounds: &Bounds3f, - render_from_medium: &Transform, - sigma_a: &Spectrum, - sigma_s: &Spectrum, - sigma_scale: Float, - g: Float, - density_grid: SampledGrid, - temperature_grid: SampledGrid, - le: &Spectrum, - le_scale: SampledGrid, - ) -> Self { - let mut sigma_a_spec = DenselySampledSpectrum::from_spectrum(sigma_a); - let mut sigma_s_spec = DenselySampledSpectrum::from_spectrum(sigma_s); - let le_spec = DenselySampledSpectrum::from_spectrum(le); - sigma_a_spec.scale(sigma_scale); - sigma_s_spec.scale(sigma_scale); - - let mut majorant_grid = MajorantGrid::new(*bounds, Point3i::new(16, 16, 16)); - let is_emissive = if temperature_grid.is_some() { - true - } else { - le_spec.max_value() > 0. - }; - - for z in 0..majorant_grid.res.z() { - for y in 0..majorant_grid.res.y() { - for x in 0..majorant_grid.res.x() { - let bounds = majorant_grid.voxel_bounds(x, y, z); - majorant_grid.set(x, y, z, density_grid.max_value(bounds)); - } - } - } - - Self { - bounds: *bounds, - render_from_medium: *render_from_medium, - sigma_a_spec, - sigma_s_spec, - density_grid, - phase: HGPhaseFunction::new(g), - temperature_grid, - le_spec, - le_scale, - is_emissive, - majorant_grid, - } - } + pub bounds: Bounds3f, + pub render_from_medium: Transform, + pub sigma_a_spec: DenselySampledSpectrum, + pub sigma_s_spec: DenselySampledSpectrum, + pub density_grid: SampledGrid, + pub phase: HGPhaseFunction, + pub temperature_grid: SampledGrid, + pub le_spec: DenselySampledSpectrum, + pub le_scale: SampledGrid, + pub is_emissive: bool, + pub majorant_grid: MajorantGrid, } impl MediumTrait for GridMedium { @@ -603,12 +528,11 @@ impl MediumTrait for GridMedium { }; let le = if scale > 0.0 { - let raw_emission = match &self.temperature_grid { - Some(grid) => { - let temp = grid.lookup(p); - BlackbodySpectrum::new(temp).sample(lambda) - } - None => self.le_spec.sample(lambda), + let raw_emission = if self.temperature_grid.is_valid() { + let temp = self.temperature_grid.lookup(p); + BlackbodySpectrum::new(temp).sample(lambda) + } else { + self.le_spec.sample(lambda) }; raw_emission * scale @@ -664,59 +588,15 @@ impl MediumTrait for GridMedium { #[repr(C)] #[derive(Debug, Clone, Copy)] pub struct RGBGridMedium { - bounds: Bounds3f, - render_from_medium: Transform, - phase: HGPhaseFunction, - le_scale: Float, - sigma_scale: Float, - sigma_a_grid: SampledGrid, - sigma_s_grid: SampledGrid, - le_grid: SampledGrid, - majorant_grid: MajorantGrid, -} - -impl RGBGridMedium { - #[allow(clippy::too_many_arguments)] - #[cfg(not(target_os = "cuda"))] - pub fn new( - bounds: &Bounds3f, - render_from_medium: &Transform, - g: Float, - sigma_a_grid: SampledGrid, - sigma_s_grid: SampledGrid, - sigma_scale: Float, - le_grid: SampledGrid, - le_scale: Float, - ) -> Self { - let mut majorant_grid = MajorantGrid::new(*bounds, Point3i::new(16, 16, 16)); - for z in 0..majorant_grid.res.x() { - for y in 0..majorant_grid.res.y() { - for x in 0..majorant_grid.res.x() { - let bounds = majorant_grid.voxel_bounds(x, y, z); - let convert = |s: &RGBUnboundedSpectrum| s.max_value(); - let max_sigma_t = sigma_a_grid - .as_ref() - .map_or(1.0, |g| g.max_value_convert(bounds, convert)) - + sigma_s_grid - .as_ref() - .map_or(1.0, |g| g.max_value_convert(bounds, convert)); - majorant_grid.set(x, y, z, sigma_scale * max_sigma_t); - } - } - } - - Self { - bounds: *bounds, - render_from_medium: *render_from_medium, - le_grid, - le_scale, - phase: HGPhaseFunction::new(g), - sigma_a_grid, - sigma_s_grid, - sigma_scale, - majorant_grid, - } - } + pub bounds: Bounds3f, + pub render_from_medium: Transform, + pub phase: HGPhaseFunction, + pub le_scale: Float, + pub sigma_scale: Float, + pub sigma_a_grid: SampledGrid, + pub sigma_s_grid: SampledGrid, + pub le_grid: SampledGrid, + pub majorant_grid: MajorantGrid, } impl MediumTrait for RGBGridMedium { @@ -789,7 +669,8 @@ impl MediumTrait for RGBGridMedium { } } -#[derive(Debug, Clone)] +#[repr(C)] +#[derive(Debug, Clone, Copy)] pub struct CloudMedium; impl MediumTrait for CloudMedium { fn is_emissive(&self) -> bool { @@ -807,7 +688,9 @@ impl MediumTrait for CloudMedium { todo!() } } -#[derive(Debug, Clone)] + +#[repr(C)] +#[derive(Debug, Clone, Copy)] pub struct NanoVDBMedium; impl MediumTrait for NanoVDBMedium { fn is_emissive(&self) -> bool { @@ -829,8 +712,8 @@ impl MediumTrait for NanoVDBMedium { #[repr(C)] #[derive(Debug, Copy, Clone)] pub struct MediumInterface { - pub inside: DevicePtr, - pub outside: DevicePtr, + pub inside: Ptr, + pub outside: Ptr, } unsafe impl Send for MediumInterface {} @@ -839,8 +722,8 @@ unsafe impl Sync for MediumInterface {} impl Default for MediumInterface { fn default() -> Self { Self { - inside: DevicePtr::null(), - outside: DevicePtr::null(), + inside: Ptr::null(), + outside: Ptr::null(), } } } @@ -848,8 +731,8 @@ impl Default for MediumInterface { impl From for MediumInterface { fn from(medium: Medium) -> Self { Self { - inside: DevicePtr::from(&medium), - outside: DevicePtr::from(&medium), + inside: Ptr::from(&medium), + outside: Ptr::from(&medium), } } } @@ -863,8 +746,8 @@ impl From<&Medium> for MediumInterface { impl MediumInterface { pub fn new(inside: &Medium, outside: &Medium) -> Self { Self { - inside: DevicePtr::from(inside), - outside: DevicePtr::from(outside), + inside: Ptr::from(inside), + outside: Ptr::from(outside), } } @@ -873,6 +756,6 @@ impl MediumInterface { } pub fn is_medium_transition(&self) -> bool { - self.inside.0 != self.outside.0 + self.inside != self.outside } } diff --git a/shared/src/core/primitive.rs b/shared/src/core/primitive.rs index bfd1928..33f1b4d 100644 --- a/shared/src/core/primitive.rs +++ b/shared/src/core/primitive.rs @@ -1,4 +1,3 @@ -use crate::core::aggregates::LinearBVHNode; use crate::core::geometry::{Bounds3f, Ray}; use crate::core::interaction::{Interaction, InteractionTrait, SurfaceInteraction}; use crate::core::light::Light; @@ -24,11 +23,11 @@ pub trait PrimitiveTrait { #[repr(C)] #[derive(Debug, Clone, Copy)] pub struct GeometricPrimitive { - shape: *const Shape, - material: *const Material, - area_light: *const Light, + shape: Ptr, + material: Ptr, + area_light: Ptr, medium_interface: MediumInterface, - alpha: *const GPUFloatTexture, + alpha: Ptr, } unsafe impl Send for GeometricPrimitive {} @@ -41,7 +40,8 @@ impl PrimitiveTrait for GeometricPrimitive { fn intersect(&self, r: &Ray, t_max: Option) -> Option { let mut si = self.shape.intersect(r, t_max)?; - if let Some(ref alpha) = self.alpha { + if !self.alpha.is_null() { + let alpha = unsafe { &self.alpha.as_ref() }; let ctx = TextureEvalContext::from(&si.intr); let a = alpha.evaluate(&ctx); if a < 1.0 { @@ -66,18 +66,21 @@ impl PrimitiveTrait for GeometricPrimitive { } } + if r.medium.is_null() { + return None; + } si.set_intersection_properties( - self.material.clone(), - self.area_light.clone(), - Some(self.medium_interface.clone()), - Some(r.medium.clone().expect("Medium not set")), + self.material, + self.area_light, + self.medium_interface.clone(), + r.medium, ); Some(si) } fn intersect_p(&self, r: &Ray, t_max: Option) -> bool { - if self.alpha.is_some() { + if !self.alpha.is_null() { self.intersect(r, t_max).is_some() } else { self.shape.intersect_p(r, t_max) @@ -160,32 +163,38 @@ impl PrimitiveTrait for AnimatedPrimitive { } } +#[repr(C)] +#[derive(Debug, Clone, Copy)] +pub struct LinearBVHNode { + bounds: Bounds3f, +} + #[repr(C)] #[derive(Debug, Clone, Copy)] pub struct BVHAggregatePrimitive { max_prims_in_node: u32, primitives: *const Ptr, - nodes: *const LinearBVHNode, + nodes: Ptr, } impl PrimitiveTrait for BVHAggregatePrimitive { fn bounds(&self) -> Bounds3f { - if !self.nodes.is_empty() { - self.nodes[0].bounds + if !self.nodes.is_null() { + self.nodes.bounds } else { Bounds3f::default() } } fn intersect(&self, r: &Ray, t_max: Option) -> Option { - if self.nodes.is_empty() { + if !self.nodes.is_null() { return None; } self.intersect(r, t_max) } fn intersect_p(&self, r: &Ray, t_max: Option) -> bool { - if self.nodes.is_empty() { + if !self.nodes.is_null() { return false; } self.intersect_p(r, t_max) diff --git a/shared/src/core/sampler.rs b/shared/src/core/sampler.rs index 89cc6e7..80e3d25 100644 --- a/shared/src/core/sampler.rs +++ b/shared/src/core/sampler.rs @@ -6,10 +6,9 @@ use crate::utils::Ptr; use crate::utils::containers::Array2D; use crate::utils::math::{ BinaryPermuteScrambler, DigitPermutation, FastOwenScrambler, NoRandomizer, OwenScrambler, - PRIME_TABLE_SIZE, Scrambler, clamp, compute_radical_inverse_permutations, encode_morton_2, - inverse_radical_inverse, lerp, log2_int, owen_scrambled_radical_inverse, permutation_element, - radical_inverse, round_up_pow2, scrambled_radical_inverse, sobol_interval_to_index, - sobol_sample, + PRIME_TABLE_SIZE, Scrambler, clamp, encode_morton_2, inverse_radical_inverse, lerp, log2_int, + owen_scrambled_radical_inverse, permutation_element, radical_inverse, round_up_pow2, + scrambled_radical_inverse, sobol_interval_to_index, sobol_sample, }; use crate::utils::rng::Rng; use crate::utils::sobol::N_SOBOL_DIMENSIONS; @@ -107,62 +106,59 @@ pub struct HaltonSampler { } impl HaltonSampler { - pub fn new( - samples_per_pixel: u32, - full_res: Point2i, - randomize: RandomizeStrategy, - seed: u64, - ) -> Self { - let digit_permutations = compute_radical_inverse_permutations(seed); - let mut base_scales = [0u64; 2]; - let mut base_exponents = [0u64; 2]; - let bases = [2, 3]; - let res_coords = [full_res.x(), full_res.y()]; - - for i in 0..2 { - let base = bases[i] as u64; - let mut scale = 1u64; - let mut exp = 0u64; - - let limit = std::cmp::min(res_coords[i], MAX_HALTON_RESOLUTION) as u64; - - while scale < limit { - scale *= base; - exp += 1; - } - - base_scales[i] = scale; - base_exponents[i] = exp; - } - - let mut mult_inverse = [0u64; 2]; - - mult_inverse[0] = - Self::multiplicative_inverse(base_scales[0] as i64, base_scales[0] as i64); - mult_inverse[1] = - Self::multiplicative_inverse(base_scales[1] as i64, base_scales[1] as i64); - - Self { - samples_per_pixel, - randomize, - digit_permutations, - base_scales, - base_exponents, - mult_inverse, - halton_index: 0, - dim: 0, - } - } + // pub fn new( + // samples_per_pixel: u32, + // full_res: Point2i, + // randomize: RandomizeStrategy, + // seed: u64, + // ) -> Self { + // let digit_permutations = compute_radical_inverse_permutations(seed); + // let mut base_scales = [0u64; 2]; + // let mut base_exponents = [0u64; 2]; + // let bases = [2, 3]; + // let res_coords = [full_res.x(), full_res.y()]; + // + // for i in 0..2 { + // let base = bases[i] as u64; + // let mut scale = 1u64; + // let mut exp = 0u64; + // + // let limit = std::cmp::min(res_coords[i], MAX_HALTON_RESOLUTION) as u64; + // + // while scale < limit { + // scale *= base; + // exp += 1; + // } + // + // base_scales[i] = scale; + // base_exponents[i] = exp; + // } + // + // let mut mult_inverse = [0u64; 2]; + // + // mult_inverse[0] = + // Self::multiplicative_inverse(base_scales[0] as i64, base_scales[0] as i64); + // mult_inverse[1] = + // Self::multiplicative_inverse(base_scales[1] as i64, base_scales[1] as i64); + // + // Self { + // samples_per_pixel, + // randomize, + // digit_permutations, + // base_scales, + // base_exponents, + // mult_inverse, + // halton_index: 0, + // dim: 0, + // } + // } fn sample_dimension(&self, dimension: u32) -> Float { if self.randomize == RandomizeStrategy::None { radical_inverse(dimension, self.halton_index) } else if self.randomize == RandomizeStrategy::PermuteDigits { - scrambled_radical_inverse( - dimension, - self.halton_index, - &self.digit_permutations[dimension as usize], - ) + let digit_perm = unsafe { &*self.digit_permutations.add(dimension as usize) }; + scrambled_radical_inverse(dimension, self.halton_index, digit_perm) } else { owen_scrambled_radical_inverse( dimension, @@ -418,13 +414,13 @@ impl SamplerTrait for PaddedSobolSampler { self.dim as u64, self.seed, ]; - let hash = hash_buffer(&hash_input, 0) as u32; + let hash = hash_buffer(&hash_input, 0); let index = permutation_element( self.sample_index as u32, self.samples_per_pixel as u32, - hash, + hash as u32, ); - self.sample_dimension(0, index, hash >> 32) + self.sample_dimension(0, index, (hash >> 32) as u32) } fn get2d(&mut self) -> Point2f { let hash_input = [ @@ -433,16 +429,16 @@ impl SamplerTrait for PaddedSobolSampler { self.dim as u64, self.seed, ]; - let hash = hash_buffer(&hash_input, 0) as u32; + let hash = hash_buffer(&hash_input, 0); let index = permutation_element( self.sample_index as u32, self.samples_per_pixel as u32, - hash, + hash as u32, ); self.dim += 2; Point2f::new( - self.sample_dimension(0, index, hash), - self.sample_dimension(1, index, hash >> 32), + self.sample_dimension(0, index, hash as u32), + self.sample_dimension(1, index, (hash >> 32) as u32), ) } @@ -630,7 +626,7 @@ impl ZSobolSampler { let mix_input = higher_digits ^ (0x55555555 * self.dim as u64); let p = (mix_bits(mix_input) >> 24) % 24; - digit = PERMUTATIONS[p as u32][digit as u32] as u64; + digit = PERMUTATIONS[p as usize][digit as usize] as u64; sample_index |= digit << digit_shift; } diff --git a/shared/src/core/shape.rs b/shared/src/core/shape.rs index ad60399..af160f0 100644 --- a/shared/src/core/shape.rs +++ b/shared/src/core/shape.rs @@ -1,6 +1,6 @@ use crate::core::geometry::{ Bounds3f, DirectionCone, Normal3f, Point2f, Point3f, Point3fi, Ray, Vector2f, Vector3f, - Vector3fi, VectorLike, + Vector3fi, VectorLike, ray, }; use crate::core::interaction::{ Interaction, InteractionTrait, MediumInteraction, SurfaceInteraction, @@ -10,7 +10,7 @@ use crate::core::material::Material; use crate::core::medium::{Medium, MediumInterface}; use crate::shapes::*; use crate::utils::math::{next_float_down, next_float_up}; -use crate::utils::{DevicePtr, Transform}; +use crate::utils::{Ptr, Transform}; use crate::{Float, PI}; use enum_dispatch::enum_dispatch; @@ -37,13 +37,13 @@ impl ShapeIntersection { pub fn set_intersection_properties( &mut self, - mtl: &Material, - area: &Light, + mtl: Ptr, + area: Ptr, prim_medium_interface: MediumInterface, - ray_medium: &Medium, + ray_medium: Ptr, ) { self.intr - .set_intersection_properties(mtl, area, ray_medium, prim_medium_interface); + .set_intersection_properties(&mtl, &area, &ray_medium, prim_medium_interface); } } @@ -118,12 +118,7 @@ impl ShapeSampleContext { } pub fn spawn_ray(&self, w: Vector3f) -> Ray { - Ray::new( - self.offset_ray_origin(w), - w, - Some(self.time), - &DevicePtr::null(), - ) + Ray::new(self.offset_ray_origin(w), w, Some(self.time), &Ptr::null()) } } diff --git a/shared/src/core/spectrum.rs b/shared/src/core/spectrum.rs index 5a9522f..ce9f39e 100644 --- a/shared/src/core/spectrum.rs +++ b/shared/src/core/spectrum.rs @@ -1,8 +1,9 @@ use crate::Float; use crate::core::color::{RGB, XYZ}; -use crate::spectra::*; use enum_dispatch::enum_dispatch; +pub use crate::spectra::*; + #[enum_dispatch] pub trait SpectrumTrait: Copy { fn evaluate(&self, lambda: Float) -> Float; diff --git a/shared/src/core/texture.rs b/shared/src/core/texture.rs index 11e9076..38db344 100644 --- a/shared/src/core/texture.rs +++ b/shared/src/core/texture.rs @@ -8,13 +8,15 @@ use crate::spectra::{ RGBAlbedoSpectrum, RGBIlluminantSpectrum, RGBUnboundedSpectrum, SampledSpectrum, SampledWavelengths, }; -use crate::textures::*; -use crate::utils::DevicePtr; + +use crate::utils::Ptr; use crate::utils::Transform; use crate::utils::math::square; use crate::{Float, INV_2_PI, INV_PI, PI}; use enum_dispatch::enum_dispatch; +pub use crate::textures::*; + #[repr(C)] #[derive(Clone, Debug, Copy)] pub struct TexCoord2D { @@ -428,8 +430,8 @@ pub trait TextureEvaluator: Send + Sync { fn can_evaluate( &self, - _ftex: &[DevicePtr], - _stex: &[DevicePtr], + _ftex: &[Ptr], + _stex: &[Ptr], ) -> bool; } @@ -453,8 +455,8 @@ impl TextureEvaluator for UniversalTextureEvaluator { fn can_evaluate( &self, - _float_textures: &[DevicePtr], - _spectrum_textures: &[DevicePtr], + _float_textures: &[Ptr], + _spectrum_textures: &[Ptr], ) -> bool { true } diff --git a/shared/src/filters/gaussian.rs b/shared/src/filters/gaussian.rs index 98d5d5d..9d3997e 100644 --- a/shared/src/filters/gaussian.rs +++ b/shared/src/filters/gaussian.rs @@ -13,27 +13,6 @@ pub struct GaussianFilter { pub sampler: FilterSampler, } -impl GaussianFilter { - pub fn new(radius: Vector2f, sigma: Float) -> Self { - let exp_x = gaussian(radius.x(), 0., sigma); - let exp_y = gaussian(radius.y(), 0., sigma); - - let sampler = FilterSampler::new(radius, move |p: Point2f| { - let gx = (gaussian(p.x(), 0., sigma) - exp_x).max(0.0); - let gy = (gaussian(p.y(), 0., sigma) - exp_y).max(0.0); - gx * gy - }); - - Self { - radius, - sigma, - exp_x: gaussian(radius.x(), 0., sigma), - exp_y: gaussian(radius.y(), 0., sigma), - sampler, - } - } -} - impl FilterTrait for GaussianFilter { fn radius(&self) -> Vector2f { self.radius diff --git a/shared/src/filters/lanczos.rs b/shared/src/filters/lanczos.rs index f962c8d..710b252 100644 --- a/shared/src/filters/lanczos.rs +++ b/shared/src/filters/lanczos.rs @@ -12,20 +12,6 @@ pub struct LanczosSincFilter { pub sampler: FilterSampler, } -impl LanczosSincFilter { - pub fn new(radius: Vector2f, tau: Float) -> Self { - let sampler = FilterSampler::new(radius, move |p: Point2f| { - windowed_sinc(p.x(), radius.x(), tau) * windowed_sinc(p.y(), radius.y(), tau) - }); - - Self { - radius, - tau, - sampler, - } - } -} - impl FilterTrait for LanczosSincFilter { fn radius(&self) -> Vector2f { self.radius diff --git a/shared/src/filters/mitchell.rs b/shared/src/filters/mitchell.rs index 6526247..4ee08db 100644 --- a/shared/src/filters/mitchell.rs +++ b/shared/src/filters/mitchell.rs @@ -12,22 +12,7 @@ pub struct MitchellFilter { } impl MitchellFilter { - pub fn new(radius: Vector2f, b: Float, c: Float) -> Self { - let sampler = FilterSampler::new(radius, move |p: Point2f| { - let nx = 2.0 * p.x() / radius.x(); - let ny = 2.0 * p.y() / radius.y(); - Self::mitchell_1d_eval(b, c, nx) * Self::mitchell_1d_eval(b, c, ny) - }); - - Self { - radius, - b, - c, - sampler, - } - } - - fn mitchell_1d_eval(b: Float, c: Float, x: Float) -> Float { + pub fn mitchell_1d_eval(b: Float, c: Float, x: Float) -> Float { let x = x.abs(); if x <= 1.0 { ((12.0 - 9.0 * b - 6.0 * c) * x.powi(3) diff --git a/shared/src/lights/diffuse.rs b/shared/src/lights/diffuse.rs index 270ec8c..6f9d22a 100644 --- a/shared/src/lights/diffuse.rs +++ b/shared/src/lights/diffuse.rs @@ -125,10 +125,8 @@ impl LightTrait for DiffuseAreaLight { rgb[c] = self.image.bilerp_channel(uv, c as i32); } - let spec = RGBIlluminantSpectrum::new( - self.image_color_space.as_ref().unwrap(), - rgb.clamp_zero(), - ); + let cs_ref = unsafe { self.colorspace.as_ref() }; + let spec = RGBIlluminantSpectrum::new(cs_ref, rgb.clamp_zero()); self.scale * spec.sample(lambda) } else { @@ -150,11 +148,9 @@ impl LightTrait for DiffuseAreaLight { for c in 0..3 { rgb[c] = self.image.get_channel(Point2i::new(x, y), c as i32); } - l += RGBIlluminantSpectrum::new( - self.image_color_space.as_ref().unwrap(), - rgb.clamp_zero(), - ) - .sample(&lambda); + + let cs_ref = unsafe { self.colorspace.as_ref() }; + l += RGBIlluminantSpectrum::new(cs_ref, rgb.clamp_zero()).sample(&lambda); } } l *= self.scale / (self.image.resolution().x() * self.image.resolution().y()) as Float; diff --git a/shared/src/lights/infinite.rs b/shared/src/lights/infinite.rs index 8595cf7..4ce555a 100644 --- a/shared/src/lights/infinite.rs +++ b/shared/src/lights/infinite.rs @@ -214,8 +214,8 @@ impl LightTrait for ImageInfiniteLight { #[cfg(not(target_os = "cuda"))] fn phi(&self, lambda: SampledWavelengths) -> SampledSpectrum { let mut sum_l = SampledSpectrum::new(0.); - let width = self.image.resolution.x(); - let height = self.image.resolution.y(); + let width = self.image.resolution().x(); + let height = self.image.resolution().y(); for v in 0..height { for u in 0..width { let mut rgb = RGB::default(); diff --git a/shared/src/lights/projection.rs b/shared/src/lights/projection.rs index 9b0b067..ef08a9a 100644 --- a/shared/src/lights/projection.rs +++ b/shared/src/lights/projection.rs @@ -3,7 +3,7 @@ use crate::core::color::RGB; use crate::core::geometry::{ Bounds2f, Bounds3f, Normal3f, Point2f, Point2i, Point3f, Ray, Vector3f, VectorLike, cos_theta, }; -use crate::core::image::DeviceImage; +use crate::core::image::{DeviceImage, ImageAccess}; use crate::core::light::{ LightBase, LightBounds, LightLiSample, LightSampleContext, LightTrait, LightType, }; @@ -91,11 +91,12 @@ impl LightTrait for ProjectionLight { fn phi(&self, lambda: SampledWavelengths) -> SampledSpectrum { let mut sum = SampledSpectrum::new(0.); - for y in 0..self.image.resolution.y() { - for x in 0..self.image.resolution.x() { + let res = self.image.resolution(); + for y in 0..res.y() { + for x in 0..res.x() { let ps = self.screen_bounds.lerp(Point2f::new( - (x as Float + 0.5) / self.image.resolution.x() as Float, - (y as Float + 0.5) / self.image.resolution.y() as Float, + (x as Float + 0.5) / res.x() as Float, + (y as Float + 0.5) / res.y() as Float, )); let w_raw = Vector3f::from(self.light_from_screen.apply_to_point(Point3f::new( ps.x(), @@ -113,7 +114,7 @@ impl LightTrait for ProjectionLight { sum += s.sample(&lambda) * dwda; } } - self.scale * self.a * sum / (self.image.resolution.x() * self.image.resolution.y()) as Float + self.scale * self.a * sum / (res.x() * res.y()) as Float } fn preprocess(&mut self, _scene_bounds: &Bounds3f) { diff --git a/shared/src/lights/sampler.rs b/shared/src/lights/sampler.rs index deeceaa..a25ce8b 100644 --- a/shared/src/lights/sampler.rs +++ b/shared/src/lights/sampler.rs @@ -6,7 +6,7 @@ use crate::core::light::{LightBounds, LightSampleContext}; use crate::spectra::{SampledSpectrum, SampledWavelengths}; use crate::utils::math::{clamp, lerp, sample_discrete}; use crate::utils::math::{safe_sqrt, square}; -use crate::utils::ptr::{DevicePtr, Slice}; +use crate::utils::ptr::Ptr; use crate::utils::sampling::AliasTable; use crate::{Float, ONE_MINUS_EPSILON, PI}; use enum_dispatch::enum_dispatch; @@ -154,14 +154,14 @@ impl CompactLightBounds { #[derive(Debug, Clone)] pub struct SampledLight { - pub light: DevicePtr, + pub light: Ptr, pub p: Float, } impl SampledLight { pub fn new(light: Light, p: Float) -> Self { Self { - light: DevicePtr::from(&light), + light: Ptr::from(&light), p, } } @@ -214,7 +214,7 @@ impl LightSamplerTrait for UniformLightSampler { let light_index = (u as u32 * self.lights_len).min(self.lights_len - 1) as usize; Some(SampledLight { - light: DevicePtr::from(&self.light(light_index)), + light: Ptr::from(&self.light(light_index)), p: 1. / self.lights_len as Float, }) } @@ -236,7 +236,7 @@ pub struct Alias { #[repr(C)] #[derive(Clone, Debug, Copy)] pub struct PowerLightSampler { - pub lights: Slice, + pub lights: Ptr, pub lights_len: u32, pub alias_table: AliasTable, } @@ -260,28 +260,24 @@ impl LightSamplerTrait for PowerLightSampler { let (light_index, pmf, _) = self.alias_table.sample(u); - let light_ref = &self.lights[light_index as usize]; + let light_ref = unsafe { self.lights.add(light_index as usize) }; Some(SampledLight { - light: DevicePtr::from(light_ref), + light: light_ref, p: pmf, }) } fn pmf(&self, light: &Light) -> Float { - if self.lights_len == 0 { - return 0.0; + let array_start = self.lights.as_raw(); + let target = light as *const Light as *mut Light; + + unsafe { + let index = target.offset_from(array_start); + + if index >= 0 && index < self.lights_len as isize { + return self.alias_table.pmf(index as u32); + } } - - let light_ptr = light as *const Light; - let start = self.lights.as_ptr(); - - let end = unsafe { start.add(self.lights.len as usize) }; - - if light_ptr >= start && light_ptr < end { - let index = unsafe { light_ptr.offset_from(start) }; - return self.alias_table.pmf(index as u32); - } - 0. } } diff --git a/shared/src/materials/coated.rs b/shared/src/materials/coated.rs index 36b95f5..2b030db 100644 --- a/shared/src/materials/coated.rs +++ b/shared/src/materials/coated.rs @@ -25,38 +25,37 @@ pub struct CoatedDiffuseMaterial { pub thickness: Ptr, pub g: Ptr, pub eta: Ptr, + pub max_depth: u32, + pub n_samples: u32, pub remap_roughness: bool, - pub max_depth: usize, - pub n_samples: usize, } impl CoatedDiffuseMaterial { #[allow(clippy::too_many_arguments)] - #[cfg(not(target_os = "cuda"))] pub fn new( - reflectance: &GPUSpectrumTexture, - u_roughness: &GPUFloatTexture, - v_roughness: &GPUFloatTexture, - thickness: &GPUFloatTexture, - albedo: &GPUSpectrumTexture, - g: &GPUFloatTexture, - eta: &Spectrum, - displacement: &GPUFloatTexture, - normal_map: &DeviceImage, + reflectance: Ptr, + u_roughness: Ptr, + v_roughness: Ptr, + thickness: Ptr, + albedo: Ptr, + g: Ptr, + eta: Ptr, + displacement: Ptr, + normal_map: Ptr, remap_roughness: bool, - max_depth: usize, - n_samples: usize, + max_depth: u32, + n_samples: u32, ) -> Self { Self { - 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), + displacement, + normal_map, + reflectance, + albedo, + u_roughness, + v_roughness, + thickness, + g, + eta, remap_roughness, max_depth, n_samples, @@ -163,46 +162,45 @@ pub struct CoatedConductorMaterial { conductor_eta: Ptr, k: Ptr, reflectance: Ptr, - remap_roughness: bool, max_depth: u32, n_samples: u32, + remap_roughness: bool, } impl CoatedConductorMaterial { #[allow(clippy::too_many_arguments)] - #[cfg(not(target_os = "cuda"))] pub fn new( - normal_map: &DeviceImage, - displacement: &GPUFloatTexture, - interface_uroughness: &GPUFloatTexture, - interface_vroughness: &GPUFloatTexture, - thickness: &GPUFloatTexture, - interface_eta: &Spectrum, - g: &GPUFloatTexture, - albedo: &GPUSpectrumTexture, - conductor_uroughness: &GPUFloatTexture, - conductor_vroughness: &GPUFloatTexture, - conductor_eta: &GPUSpectrumTexture, - k: &GPUSpectrumTexture, - reflectance: &GPUSpectrumTexture, - remap_roughness: bool, + 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, max_depth: u32, n_samples: u32, + remap_roughness: bool, ) -> Self { Self { - 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), + displacement, + normal_map, + interface_uroughness, + interface_vroughness, + thickness, + interface_eta, + g, + albedo, + conductor_uroughness, + conductor_vroughness, + conductor_eta, + k, + reflectance, remap_roughness, max_depth, n_samples, @@ -238,12 +236,9 @@ impl MaterialTrait for CoatedConductorMaterial { } let (mut ce, mut ck) = if !self.conductor_eta.is_null() { - let k_tex = self - .k - .as_ref() - .expect("CoatedConductor: 'k' must be provided if 'conductor_eta' is present"); + let k_tex = self.k; let ce = tex_eval.evaluate_spectrum(&self.conductor_eta, ctx, lambda); - let ck = tex_eval.evaluate_spectrum(k_tex, ctx, lambda); + let ck = tex_eval.evaluate_spectrum(unsafe { k_tex.as_ref() }, ctx, lambda); (ce, ck) } else { let r = SampledSpectrum::clamp( @@ -282,8 +277,8 @@ impl MaterialTrait for CoatedConductorMaterial { thick, a, gg, - self.max_depth as usize, - self.n_samples as usize, + self.max_depth, + self.n_samples, )); BSDF::new(ctx.ns, ctx.dpdus, Ptr::from(&bxdf)) } diff --git a/shared/src/materials/complex.rs b/shared/src/materials/complex.rs index 974eb5e..9761cd0 100644 --- a/shared/src/materials/complex.rs +++ b/shared/src/materials/complex.rs @@ -13,33 +13,33 @@ use crate::core::spectrum::{Spectrum, SpectrumTrait}; use crate::core::texture::{GPUFloatTexture, GPUSpectrumTexture, TextureEvaluator}; use crate::spectra::{SampledSpectrum, SampledWavelengths}; use crate::textures::GPUSpectrumMixTexture; -use crate::utils::DevicePtr; +use crate::utils::Ptr; use crate::utils::math::clamp; #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct HairMaterial { - pub sigma_a: DevicePtr, - pub color: DevicePtr, - pub eumelanin: DevicePtr, - pub pheomelanin: DevicePtr, - pub eta: DevicePtr, - pub beta_m: DevicePtr, - pub beta_n: DevicePtr, - pub alpha: DevicePtr, + pub sigma_a: Ptr, + pub color: Ptr, + pub eumelanin: Ptr, + pub pheomelanin: Ptr, + pub eta: Ptr, + pub beta_m: Ptr, + pub beta_n: Ptr, + pub alpha: Ptr, } impl HairMaterial { #[cfg(not(target_os = "cuda"))] pub fn new( - sigma_a: DevicePtr, - color: DevicePtr, - eumelanin: DevicePtr, - pheomelanin: DevicePtr, - eta: DevicePtr, - beta_m: DevicePtr, - beta_n: DevicePtr, - alpha: DevicePtr, + sigma_a: Ptr, + color: Ptr, + eumelanin: Ptr, + pheomelanin: Ptr, + eta: Ptr, + beta_m: Ptr, + beta_n: Ptr, + alpha: Ptr, ) -> Self { Self { sigma_a, @@ -80,8 +80,8 @@ impl MaterialTrait for HairMaterial { todo!() } - fn get_displacement(&self) -> DevicePtr { - DevicePtr::null() + fn get_displacement(&self) -> Ptr { + Ptr::null() } fn has_subsurface_scattering(&self) -> bool { @@ -92,9 +92,9 @@ impl MaterialTrait for HairMaterial { #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct MeasuredMaterial { - pub displacement: DevicePtr, - pub normal_map: DevicePtr, - pub brdf: DevicePtr, + pub displacement: Ptr, + pub normal_map: Ptr, + pub brdf: Ptr, } impl MaterialTrait for MeasuredMaterial { @@ -125,7 +125,7 @@ impl MaterialTrait for MeasuredMaterial { Some(&*self.normal_map) } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { self.displacement } @@ -137,16 +137,16 @@ impl MaterialTrait for MeasuredMaterial { #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct SubsurfaceMaterial { - pub normal_map: DevicePtr, - pub displacement: DevicePtr, - pub sigma_a: DevicePtr, - pub sigma_s: DevicePtr, - pub reflectance: DevicePtr, - pub mfp: DevicePtr, + pub normal_map: Ptr, + pub displacement: Ptr, + pub sigma_a: Ptr, + pub sigma_s: Ptr, + pub reflectance: Ptr, + pub mfp: Ptr, pub eta: Float, pub scale: Float, - pub u_roughness: DevicePtr, - pub v_roughness: DevicePtr, + pub u_roughness: Ptr, + pub v_roughness: Ptr, pub remap_roughness: bool, pub table: BSSRDFTable, } @@ -177,7 +177,7 @@ impl MaterialTrait for SubsurfaceMaterial { todo!() } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { todo!() } diff --git a/shared/src/materials/dielectric.rs b/shared/src/materials/dielectric.rs index 99f62ef..2088cbe 100644 --- a/shared/src/materials/dielectric.rs +++ b/shared/src/materials/dielectric.rs @@ -10,17 +10,17 @@ 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 DielectricMaterial { - normal_map: DevicePtr, - displacement: DevicePtr, - u_roughness: DevicePtr, - v_roughness: DevicePtr, - eta: DevicePtr, + normal_map: Ptr, + displacement: Ptr, + u_roughness: Ptr, + v_roughness: Ptr, + eta: Ptr, remap_roughness: bool, } @@ -51,7 +51,7 @@ impl MaterialTrait for DielectricMaterial { let distrib = TrowbridgeReitzDistribution::new(u_rough, v_rough); let bxdf = BxDF::Dielectric(DielectricBxDF::new(sampled_eta, distrib)); - BSDF::new(ctx.ns, ctx.dpdus, DevicePtr::from(&bxdf)) + BSDF::new(ctx.ns, ctx.dpdus, Ptr::from(&bxdf)) } fn get_bssrdf( @@ -71,7 +71,7 @@ impl MaterialTrait for DielectricMaterial { Some(&*self.normal_map) } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { self.displacement } @@ -83,9 +83,9 @@ impl MaterialTrait for DielectricMaterial { #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct ThinDielectricMaterial { - pub displacement: DevicePtr, - pub normal_map: DevicePtr, - pub eta: DevicePtr, + pub displacement: Ptr, + pub normal_map: Ptr, + pub eta: Ptr, } impl MaterialTrait for ThinDielectricMaterial { fn get_bsdf( @@ -113,7 +113,7 @@ impl MaterialTrait for ThinDielectricMaterial { Some(&*self.normal_map) } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { self.displacement } diff --git a/shared/src/materials/diffuse.rs b/shared/src/materials/diffuse.rs index 40ce86f..0e467ca 100644 --- a/shared/src/materials/diffuse.rs +++ b/shared/src/materials/diffuse.rs @@ -11,15 +11,15 @@ 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 DiffuseMaterial { - pub normal_map: DevicePtr, - pub displacement: DevicePtr, - pub reflectance: DevicePtr, + pub normal_map: Ptr, + pub displacement: Ptr, + pub reflectance: Ptr, } impl MaterialTrait for DiffuseMaterial { @@ -31,7 +31,7 @@ impl MaterialTrait for DiffuseMaterial { ) -> BSDF { let r = tex_eval.evaluate_spectrum(&self.reflectance, ctx, lambda); let bxdf = BxDF::Diffuse(DiffuseBxDF::new(r)); - BSDF::new(ctx.ns, ctx.dpdus, DevicePtr::from(&bxdf)) + BSDF::new(ctx.ns, ctx.dpdus, Ptr::from(&bxdf)) } fn get_bssrdf( @@ -51,7 +51,7 @@ impl MaterialTrait for DiffuseMaterial { Some(&*self.normal_map) } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { self.displacement } @@ -63,10 +63,10 @@ impl MaterialTrait for DiffuseMaterial { #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct DiffuseTransmissionMaterial { - pub image: DevicePtr, - pub displacement: DevicePtr, - pub reflectance: DevicePtr, - pub transmittance: DevicePtr, + pub image: Ptr, + pub displacement: Ptr, + pub reflectance: Ptr, + pub transmittance: Ptr, pub scale: Float, } @@ -96,7 +96,7 @@ impl MaterialTrait for DiffuseTransmissionMaterial { Some(&*self.image) } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { self.displacement } diff --git a/shared/src/materials/mix.rs b/shared/src/materials/mix.rs index 135eb3c..424b62d 100644 --- a/shared/src/materials/mix.rs +++ b/shared/src/materials/mix.rs @@ -10,14 +10,14 @@ 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::hash::hash_float; use crate::utils::math::clamp; -use crate::utils::{DevicePtr, Ptr}; #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct MixMaterial { - pub amount: DevicePtr, + pub amount: Ptr, pub materials: [Ptr; 2], } @@ -73,7 +73,7 @@ impl MaterialTrait for MixMaterial { None } - fn get_displacement(&self) -> DevicePtr { + fn get_displacement(&self) -> Ptr { panic!( "MixMaterial::get_displacement() shouldn't be called. \ Displacement is not supported on Mix materials directly." diff --git a/shared/src/shapes/bilinear.rs b/shared/src/shapes/bilinear.rs index d1efb16..7cc2b94 100644 --- a/shared/src/shapes/bilinear.rs +++ b/shared/src/shapes/bilinear.rs @@ -61,11 +61,7 @@ impl BilinearPatchShape { #[inline(always)] fn get_vertex_indices(&self) -> [usize; 4] { unsafe { - let base_ptr = self - .mesh - .vertex_indices - .0 - .add((self.blp_index as usize) * 4); + let base_ptr = self.mesh.vertex_indices.add((self.blp_index as usize) * 4); [ *base_ptr.add(0) as usize, *base_ptr.add(1) as usize, @@ -80,10 +76,10 @@ impl BilinearPatchShape { let [v0, v1, v2, v3] = self.get_vertex_indices(); unsafe { [ - *self.mesh.p.0.add(v0), - *self.mesh.p.0.add(v1), - *self.mesh.p.0.add(v2), - *self.mesh.p.0.add(v3), + *self.mesh.p.add(v0), + *self.mesh.p.add(v1), + *self.mesh.p.add(v2), + *self.mesh.p.add(v3), ] } } @@ -96,10 +92,10 @@ impl BilinearPatchShape { let [v0, v1, v2, v3] = self.get_vertex_indices(); unsafe { Some([ - *self.mesh.uv.0.add(v0), - *self.mesh.uv.0.add(v1), - *self.mesh.uv.0.add(v2), - *self.mesh.uv.0.add(v3), + *self.mesh.uv.add(v0), + *self.mesh.uv.add(v1), + *self.mesh.uv.add(v2), + *self.mesh.uv.add(v3), ]) } } @@ -112,10 +108,10 @@ impl BilinearPatchShape { let [v0, v1, v2, v3] = self.get_vertex_indices(); unsafe { Some([ - *self.mesh.n.0.add(v0), - *self.mesh.n.0.add(v1), - *self.mesh.n.0.add(v2), - *self.mesh.n.0.add(v3), + *self.mesh.n.add(v0), + *self.mesh.n.add(v1), + *self.mesh.n.add(v2), + *self.mesh.n.add(v3), ]) } } diff --git a/shared/src/shapes/triangle.rs b/shared/src/shapes/triangle.rs index 593293a..6c6ed96 100644 --- a/shared/src/shapes/triangle.rs +++ b/shared/src/shapes/triangle.rs @@ -45,11 +45,7 @@ impl TriangleShape { #[inline(always)] fn get_vertex_indices(&self) -> [usize; 3] { unsafe { - let base_ptr = self - .mesh - .vertex_indices - .0 - .add((self.tri_index as usize) * 3); + let base_ptr = self.mesh.vertex_indices.add((self.tri_index as usize) * 3); [ *base_ptr.add(0) as usize, *base_ptr.add(1) as usize, @@ -63,9 +59,9 @@ impl TriangleShape { let [v0, v1, v2] = self.get_vertex_indices(); unsafe { [ - *self.mesh.p.0.add(v0), - *self.mesh.p.0.add(v1), - *self.mesh.p.0.add(v2), + *self.mesh.p.add(v0), + *self.mesh.p.add(v1), + *self.mesh.p.add(v2), ] } } @@ -78,9 +74,9 @@ impl TriangleShape { let [v0, v1, v2] = self.get_vertex_indices(); unsafe { Some([ - *self.mesh.uv.0.add(v0), - *self.mesh.uv.0.add(v1), - *self.mesh.uv.0.add(v2), + *self.mesh.uv.add(v0), + *self.mesh.uv.add(v1), + *self.mesh.uv.add(v2), ]) } } @@ -93,9 +89,9 @@ impl TriangleShape { let [v0, v1, v2] = self.get_vertex_indices(); unsafe { Some([ - *self.mesh.s.0.add(v0), - *self.mesh.s.0.add(v1), - *self.mesh.s.0.add(v2), + *self.mesh.s.add(v0), + *self.mesh.s.add(v1), + *self.mesh.s.add(v2), ]) } } @@ -108,9 +104,9 @@ impl TriangleShape { let [v0, v1, v2] = self.get_vertex_indices(); unsafe { Some([ - *self.mesh.n.0.add(v0), - *self.mesh.n.0.add(v1), - *self.mesh.n.0.add(v2), + *self.mesh.n.add(v0), + *self.mesh.n.add(v1), + *self.mesh.n.add(v2), ]) } } @@ -213,7 +209,7 @@ impl TriangleShape { ); isect.face_index = if !self.mesh.face_indices.is_null() { - unsafe { *self.mesh.face_indices.0.add(self.tri_index as usize) } + unsafe { *self.mesh.face_indices.add(self.tri_index as usize) } } else { 0 }; diff --git a/shared/src/spectra/colorspace.rs b/shared/src/spectra/colorspace.rs index a79e8d0..77dee5d 100644 --- a/shared/src/spectra/colorspace.rs +++ b/shared/src/spectra/colorspace.rs @@ -3,17 +3,17 @@ use crate::core::geometry::Point2f; use crate::core::pbrt::Float; use crate::spectra::{DenselySampledSpectrum, SampledSpectrum}; use crate::utils::math::SquareMatrix3f; -use crate::utils::ptr::DevicePtr; +use crate::utils::ptr::Ptr; use std::cmp::{Eq, PartialEq}; #[repr(C)] #[derive(Copy, Debug, Clone)] pub struct StandardColorSpaces { - pub srgb: DevicePtr, - pub dci_p3: DevicePtr, - pub rec2020: DevicePtr, - pub aces2065_1: DevicePtr, + pub srgb: Ptr, + pub dci_p3: Ptr, + pub rec2020: Ptr, + pub aces2065_1: Ptr, } #[repr(C)] @@ -24,7 +24,7 @@ pub struct RGBColorSpace { pub b: Point2f, pub w: Point2f, pub illuminant: DenselySampledSpectrum, - pub rgb_to_spectrum_table: DevicePtr, + pub rgb_to_spectrum_table: Ptr, pub xyz_from_rgb: SquareMatrix3f, pub rgb_from_xyz: SquareMatrix3f, } @@ -42,7 +42,7 @@ impl RGBColorSpace { } pub fn to_rgb_coeffs(&self, rgb: RGB) -> RGBSigmoidPolynomial { - self.rgb_to_spectrum_table.to_polynomial(rgb) + self.rgb_to_spectrum_table.evaluate(rgb) } pub fn convert_colorspace(&self, other: &RGBColorSpace) -> SquareMatrix3f { diff --git a/shared/src/spectra/rgb.rs b/shared/src/spectra/rgb.rs index 1f0d454..1044b10 100644 --- a/shared/src/spectra/rgb.rs +++ b/shared/src/spectra/rgb.rs @@ -4,7 +4,7 @@ use super::{ }; use crate::core::color::{RGB, RGBSigmoidPolynomial, XYZ}; use crate::core::spectrum::SpectrumTrait; -use crate::utils::DevicePtr; +use crate::utils::Ptr; use crate::Float; @@ -69,7 +69,7 @@ impl SpectrumTrait for UnboundedRGBSpectrum { pub struct RGBIlluminantSpectrum { pub scale: Float, pub rsp: RGBSigmoidPolynomial, - pub illuminant: DevicePtr, + pub illuminant: Ptr, } impl RGBIlluminantSpectrum { @@ -85,7 +85,7 @@ impl RGBIlluminantSpectrum { Self { scale, rsp, - illuminant: DevicePtr::from(&illuminant), + illuminant: Ptr::from(&illuminant), } } } diff --git a/shared/src/spectra/simple.rs b/shared/src/spectra/simple.rs index 9738f01..2b30fca 100644 --- a/shared/src/spectra/simple.rs +++ b/shared/src/spectra/simple.rs @@ -4,7 +4,7 @@ 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::utils::ptr::Ptr; use core::slice; use std::hash::{Hash, Hasher}; use std::sync::LazyLock; @@ -36,7 +36,7 @@ impl SpectrumTrait for ConstantSpectrum { pub struct DenselySampledSpectrum { pub lambda_min: i32, pub lambda_max: i32, - pub values: DevicePtr, + pub values: Ptr, } unsafe impl Send for DenselySampledSpectrum {} @@ -54,7 +54,7 @@ impl DenselySampledSpectrum { #[inline(always)] fn get(&self, idx: u32) -> Float { - unsafe { *self.values.0.add(idx as usize) } + unsafe { *self.values.add(idx as usize) } } } @@ -62,7 +62,7 @@ impl PartialEq for DenselySampledSpectrum { fn eq(&self, other: &Self) -> bool { self.lambda_min == other.lambda_min && self.lambda_max == other.lambda_max - && self.values.0 == other.values.0 + && self.values == other.values } } @@ -91,7 +91,7 @@ impl SpectrumTrait for DenselySampledSpectrum { s[i] = 0.0; } else { unsafe { - s[i] = *self.values.0.add(offset as usize); + s[i] = *self.values.add(offset as usize); } } } @@ -104,7 +104,7 @@ impl SpectrumTrait for DenselySampledSpectrum { if offset < 0 || offset >= n { 0.0 } else { - unsafe { *self.values.0.add(offset as usize) } + unsafe { *self.values.add(offset as usize) } } } @@ -118,7 +118,7 @@ impl SpectrumTrait for DenselySampledSpectrum { for i in 0..n { unsafe { - let val = *self.values.0.add(i); + let val = *self.values.add(i); if val > max_val { max_val = val; } @@ -131,20 +131,20 @@ impl SpectrumTrait for DenselySampledSpectrum { #[repr(C)] #[derive(Debug, Clone, Copy)] pub struct PiecewiseLinearSpectrum { - pub lambdas: DevicePtr, - pub values: DevicePtr, + pub lambdas: Ptr, + pub values: Ptr, pub count: u32, } impl PiecewiseLinearSpectrum { #[inline(always)] fn lambda(&self, i: u32) -> Float { - unsafe { *self.lambdas.0.add(i as usize) } + unsafe { *self.lambdas.add(i as usize) } } #[inline(always)] fn value(&self, i: u32) -> Float { - unsafe { *self.values.0.add(i as usize) } + unsafe { *self.values.add(i as usize) } } } @@ -186,7 +186,7 @@ impl SpectrumTrait for PiecewiseLinearSpectrum { for i in 0..n { unsafe { - let val = *self.values.0.add(i as usize); + let val = *self.values.add(i as usize); if val > max_val { max_val = val; } diff --git a/shared/src/textures/bilerp.rs b/shared/src/textures/bilerp.rs index a521053..57fd128 100644 --- a/shared/src/textures/bilerp.rs +++ b/shared/src/textures/bilerp.rs @@ -3,7 +3,7 @@ use crate::core::spectrum::Spectrum; use crate::core::spectrum::SpectrumTrait; use crate::core::texture::{TextureEvalContext, TextureMapping2D}; use crate::spectra::{SampledSpectrum, SampledWavelengths}; -use crate::utils::{DevicePtr, Transform}; +use crate::utils::{Ptr, Transform}; #[repr(C)] #[derive(Debug, Copy, Clone)] @@ -48,19 +48,19 @@ impl FloatBilerpTexture { #[derive(Clone, Copy, Debug)] pub struct SpectrumBilerpTexture { pub mapping: TextureMapping2D, - pub v00: DevicePtr, - pub v01: DevicePtr, - pub v10: DevicePtr, - pub v11: DevicePtr, + pub v00: Ptr, + pub v01: Ptr, + pub v10: Ptr, + pub v11: Ptr, } impl SpectrumBilerpTexture { pub fn new( mapping: TextureMapping2D, - v00: DevicePtr, - v01: DevicePtr, - v10: DevicePtr, - v11: DevicePtr, + v00: Ptr, + v01: Ptr, + v10: Ptr, + v11: Ptr, ) -> Self { Self { mapping, diff --git a/shared/src/textures/checkerboard.rs b/shared/src/textures/checkerboard.rs index 629ccab..2b28fe2 100644 --- a/shared/src/textures/checkerboard.rs +++ b/shared/src/textures/checkerboard.rs @@ -4,12 +4,12 @@ use crate::core::texture::{ TextureMapping3DTrait, }; use crate::spectra::{SampledSpectrum, SampledWavelengths}; -use crate::utils::{DevicePtr, Ptr, math::square}; +use crate::utils::{Ptr, math::square}; fn checkerboard( ctx: &TextureEvalContext, - map2d: DevicePtr, - map3d: DevicePtr, + map2d: Ptr, + map3d: Ptr, ) -> Float { let d = |x: Float| -> Float { let y = x / 2. - (x / 2.).floor() - 0.5; @@ -43,8 +43,8 @@ fn checkerboard( #[repr(C)] #[derive(Debug, Copy, Clone)] pub struct FloatCheckerboardTexture { - pub map2d: DevicePtr, - pub map3d: DevicePtr, + pub map2d: Ptr, + pub map3d: Ptr, pub tex: [Ptr; 2], } @@ -74,8 +74,8 @@ impl FloatCheckerboardTexture { #[repr(C)] #[derive(Clone, Copy, Debug)] pub struct SpectrumCheckerboardTexture { - pub map2d: DevicePtr, - pub map3d: DevicePtr, + pub map2d: Ptr, + pub map3d: Ptr, pub tex: [Ptr; 2], } diff --git a/shared/src/textures/dots.rs b/shared/src/textures/dots.rs index 77158eb..896466f 100644 --- a/shared/src/textures/dots.rs +++ b/shared/src/textures/dots.rs @@ -4,7 +4,7 @@ use crate::core::texture::{ GPUFloatTexture, GPUSpectrumTexture, TextureEvalContext, TextureMapping2D, }; use crate::spectra::sampled::{SampledSpectrum, SampledWavelengths}; -use crate::utils::DevicePtr; +use crate::utils::Ptr; use crate::utils::math::square; use crate::utils::noise::noise_2d; @@ -28,8 +28,8 @@ fn inside_polka_dot(st: Point2f) -> bool { #[derive(Debug, Clone, Copy)] pub struct FloatDotsTexture { pub mapping: TextureMapping2D, - pub outside_dot: DevicePtr, - pub inside_dot: DevicePtr, + pub outside_dot: Ptr, + pub inside_dot: Ptr, } impl FloatDotsTexture { @@ -53,8 +53,8 @@ impl FloatDotsTexture { #[derive(Clone, Copy, Debug)] pub struct SpectrumDotsTexture { pub mapping: TextureMapping2D, - pub outside_dot: DevicePtr, - pub inside_dot: DevicePtr, + pub outside_dot: Ptr, + pub inside_dot: Ptr, } impl SpectrumDotsTexture { diff --git a/shared/src/textures/image.rs b/shared/src/textures/image.rs index 901ef54..73770b1 100644 --- a/shared/src/textures/image.rs +++ b/shared/src/textures/image.rs @@ -23,6 +23,7 @@ pub struct GPUSpectrumImageTexture { } impl GPUSpectrumImageTexture { + #[allow(unused)] pub fn evaluate( &self, ctx: &TextureEvalContext, @@ -44,12 +45,14 @@ impl GPUSpectrumImageTexture { let d_p_dy = [c.dsdy, c.dtdy]; let tex_color = if self.is_single_channel { - let val: Float = - unsafe { intrinsics::tex2d_grad(self.tex_obj, u, v, d_p_dx, d_p_dy) }; + let val = 0.; + // let val: Float = + // unsafe { intrinsics::tex2d_grad(self.tex_obj, u, v, d_p_dx, d_p_dy) }; RGB::new(val, val, val) } else { - let val: [Float; 4] = - unsafe { intrinsics::tex2d_grad(self.tex_obj, u, v, d_p_dx, d_p_dy) }; + let val = [0., 0., 0., 0.]; + // let val: [Float; 4] = + // unsafe { intrinsics::tex2d_grad(self.tex_obj, u, v, d_p_dx, d_p_dy) }; RGB::new(val[0], val[1], val[2]) }; @@ -86,6 +89,7 @@ impl GPUFloatImageTexture { return 0.; } #[cfg(feature = "cuda")] + #[allow(unused)] { use cuda_std::intrinsics; let c = self.mapping.map(ctx); @@ -93,10 +97,12 @@ impl GPUFloatImageTexture { let v = 1.0 - c.st.y(); let d_p_dx = [c.dsdx, c.dtdx]; let d_p_dy = [c.dsdy, c.dtdy]; - let val: Float = unsafe { intrinsics::tex2d_grad(self.tex_obj, u, v, d_p_dx, d_p_dy) }; + // let val: Float = unsafe { intrinsics::tex2d_grad(self.tex_obj, u, v, d_p_dx, d_p_dy) }; + let val: Float = 0.; let result = if self.invert { - (1.0 - val).max(0.0) // Invert the pixel intensity + // Invert the pixel intensity + (1.0 - val).max(0.0) } else { val }; diff --git a/shared/src/textures/marble.rs b/shared/src/textures/marble.rs index f6f2a47..47bd726 100644 --- a/shared/src/textures/marble.rs +++ b/shared/src/textures/marble.rs @@ -6,7 +6,7 @@ use crate::core::texture::{TextureEvalContext, TextureMapping3D}; use crate::spectra::{RGBAlbedoSpectrum, RGBColorSpace, SampledSpectrum, SampledWavelengths}; use crate::utils::math::clamp; use crate::utils::noise::fbm; -use crate::utils::ptr::DevicePtr; +use crate::utils::ptr::Ptr; use crate::utils::splines::evaluate_cubic_bezier; #[repr(C)] @@ -18,7 +18,7 @@ pub struct MarbleTexture { pub scale: Float, pub variation: Float, // TODO: DO not forget to pass StandardColorSpace here!! - pub colorspace: DevicePtr, + pub colorspace: Ptr, } unsafe impl Send for MarbleTexture {} diff --git a/shared/src/textures/ptex.rs b/shared/src/textures/ptex.rs index 19abde2..e85f68a 100644 --- a/shared/src/textures/ptex.rs +++ b/shared/src/textures/ptex.rs @@ -6,7 +6,7 @@ use crate::spectra::{ RGBAlbedoSpectrum, RGBColorSpace, RGBIlluminantSpectrum, RGBUnboundedSpectrum, SampledSpectrum, SampledWavelengths, StandardColorSpaces, }; -use crate::utils::ptr::{DevicePtr, Slice}; +use crate::utils::Ptr; #[repr(C)] #[derive(Debug, Clone, Copy)] @@ -23,7 +23,7 @@ impl GPUFloatPtexTexture { #[repr(C)] #[derive(Clone, Debug, Copy)] pub struct GPUSpectrumPtexTexture { - pub face_values: Slice, + pub face_values: *const RGB, pub n_faces: u32, pub spectrum_type: SpectrumType, pub colorspaces: StandardColorSpaces, @@ -36,16 +36,16 @@ impl GPUSpectrumPtexTexture { lambda: &SampledWavelengths, ) -> SampledSpectrum { let index = ctx.face_index.clamp(0, self.n_faces.saturating_sub(1)); - let rgb = self.face_values[index as usize]; + let rgb = unsafe { &*self.face_values.add(index as usize) }; let s_rgb = self.colorspaces.srgb; match self.spectrum_type { - SpectrumType::Unbounded => RGBUnboundedSpectrum::new(&s_rgb, rgb).sample(lambda), + SpectrumType::Unbounded => RGBUnboundedSpectrum::new(&s_rgb, *rgb).sample(lambda), SpectrumType::Albedo => { let clamped_rgb = rgb.clamp(0.0, 1.0); RGBAlbedoSpectrum::new(&s_rgb, clamped_rgb).sample(lambda) } - SpectrumType::Illuminant => RGBIlluminantSpectrum::new(&s_rgb, rgb).sample(lambda), + SpectrumType::Illuminant => RGBIlluminantSpectrum::new(&s_rgb, *rgb).sample(lambda), } } } diff --git a/shared/src/utils/containers.rs b/shared/src/utils/containers.rs index 0746557..9d5ebe9 100644 --- a/shared/src/utils/containers.rs +++ b/shared/src/utils/containers.rs @@ -1,4 +1,8 @@ +use crate::core::geometry::{ + Bounds2i, Bounds3f, Bounds3i, Point2i, Point3f, Point3i, Vector2i, Vector3f, Vector3i, +}; use crate::core::pbrt::Float; +use crate::utils::Ptr; use crate::utils::math::lerp; use std::collections::HashSet; use std::collections::hash_map::RandomState; @@ -6,10 +10,6 @@ use std::hash::{BuildHasher, Hash, Hasher}; use std::ops::{Add, Index, IndexMut, Mul, Sub}; use std::sync::{Arc, Mutex, RwLock}; -use crate::core::geometry::{ - Bounds2i, Bounds3f, Bounds3i, Point2i, Point3f, Point3i, Vector2i, Vector3f, Vector3i, -}; - pub trait Interpolatable: Copy + Default + Add + Sub + Mul { @@ -120,7 +120,7 @@ impl IndexMut<(i32, i32)> for Array2D { #[repr(C)] #[derive(Debug, Clone, Copy)] pub struct SampledGrid { - pub values: *const T, + pub values: Ptr, pub values_len: u32, pub nx: i32, pub ny: i32, @@ -135,7 +135,7 @@ impl SampledGrid { pub fn new(slice: &[T], nx: i32, ny: i32, nz: i32) -> Self { assert_eq!(slice.len(), (nx * ny * nz) as usize); Self { - values: slice.as_ptr(), + values: Ptr::from(slice), values_len: (nx * ny * nz) as u32, nx, ny, @@ -145,7 +145,7 @@ impl SampledGrid { pub fn empty() -> Self { Self { - values: core::ptr::null(), + values: Ptr::null(), values_len: 0, nx: 0, ny: 0, diff --git a/shared/src/utils/error.rs b/shared/src/utils/error.rs index e3ef18b..774f00d 100644 --- a/shared/src/utils/error.rs +++ b/shared/src/utils/error.rs @@ -1,12 +1,14 @@ use std::fmt; use std::sync::Arc; -#[derive(Error, Debug)] +#[derive(Debug)] pub enum LlsError { SingularMatrix, } -#[derive(Error, Debug, Clone, PartialEq, Eq)] +impl std::error::Error for LlsError {} + +#[derive(Debug, Clone, PartialEq, Eq)] pub enum InversionError { SingularMatrix, EmptyMatrix, @@ -30,3 +32,5 @@ impl fmt::Display for InversionError { } } } + +impl std::error::Error for InversionError {} diff --git a/shared/src/utils/math.rs b/shared/src/utils/math.rs index c9c9b12..a3d45c9 100644 --- a/shared/src/utils/math.rs +++ b/shared/src/utils/math.rs @@ -5,7 +5,7 @@ use crate::core::pbrt::{Float, FloatBitOps, FloatBits, ONE_MINUS_EPSILON, PI, PI use crate::utils::hash::{hash_buffer, mix_bits}; use crate::utils::sobol::{SOBOL_MATRICES_32, VDC_SOBOL_MATRICES, VDC_SOBOL_MATRICES_INV}; -use crate::utils::DevicePtr; +use crate::utils::Ptr; use half::f16; use num_traits::{Float as NumFloat, Num, One, Signed, Zero}; use std::error::Error; @@ -647,7 +647,7 @@ pub fn round_up_pow2(mut n: i32) -> i32 { pub const PRIME_TABLE_SIZE: usize = 1000; -const PRIMES: [i32; PRIME_TABLE_SIZE] = [ +pub const PRIMES: [i32; PRIME_TABLE_SIZE] = [ 2, 3, 5, 7, 11, // Subsequent prime numbers 13, 17, 19, 23, 29, 31, 37, 41, 43, 47, 53, 59, 61, 67, 71, 73, 79, 83, 89, 97, 101, 103, 107, 109, 113, 127, 131, 137, 139, 149, 151, 157, 163, 167, 173, 179, 181, 191, 193, 197, 199, 211, @@ -753,14 +753,15 @@ pub fn inverse_radical_inverse(mut inverse: u64, base: u64, n_digits: u64) -> u6 pub struct DigitPermutation { pub base: u32, pub n_digits: u32, - pub permutations: DevicePtr, + pub permutations: Ptr, } impl DigitPermutation { #[inline(always)] pub fn permute(&self, digit_index: i32, digit_value: i32) -> i32 { let idx = (digit_index * self.base as i32 + digit_value) as usize; - self.permutations[idx] as i32 + let permutation = unsafe { *self.permutations.add(idx.into()) }; + permutation as i32 } } diff --git a/shared/src/utils/mod.rs b/shared/src/utils/mod.rs index e412715..4c1c841 100644 --- a/shared/src/utils/mod.rs +++ b/shared/src/utils/mod.rs @@ -1,5 +1,3 @@ -use core::sync::atomic::{AtomicU32, AtomicU64, Ordering}; - pub mod containers; pub mod error; pub mod hash; @@ -15,9 +13,12 @@ pub mod sobol; pub mod splines; pub mod transform; -pub use ptr::{DevicePtr, Ptr}; +pub use ptr::Ptr; pub use transform::{AnimatedTransform, Transform, TransformGeneric}; +use crate::Float; +use core::sync::atomic::{AtomicU32, AtomicU64, Ordering}; + #[inline] pub fn find_interval(sz: u32, pred: F) -> u32 where @@ -57,6 +58,8 @@ where i } +#[repr(C)] +#[derive(Debug, Default)] pub struct AtomicFloat { bits: AtomicU32, } @@ -68,11 +71,11 @@ impl AtomicFloat { } } - pub fn get(&self) -> f32 { - f32::from_bits(self.bits.load(Ordering::Relaxed)) + pub fn get(&self) -> Float { + Float::from_bits(self.bits.load(Ordering::Relaxed)) } - pub fn set(&self, val: f32) { + pub fn set(&self, val: Float) { self.bits.store(val.to_bits(), Ordering::Relaxed); } diff --git a/shared/src/utils/ptr.rs b/shared/src/utils/ptr.rs index a02e3b8..37e6a31 100644 --- a/shared/src/utils/ptr.rs +++ b/shared/src/utils/ptr.rs @@ -14,7 +14,14 @@ impl Clone for Ptr { } impl Copy for Ptr {} -// Ptr is just a pointer - Send/Sync depends on T +impl PartialEq for Ptr { + fn eq(&self, other: &Self) -> bool { + self.ptr == other.ptr + } +} + +impl Eq for Ptr {} + unsafe impl Send for Ptr {} unsafe impl Sync for Ptr {} diff --git a/shared/src/utils/sampling.rs b/shared/src/utils/sampling.rs index 94bac16..061bbd3 100644 --- a/shared/src/utils/sampling.rs +++ b/shared/src/utils/sampling.rs @@ -722,6 +722,27 @@ impl DevicePiecewiseConstant1D { self.n } + fn find_interval(&self, u: Float) -> usize { + let mut size = self.n as usize; + let mut first = 0; + + while size > 0 { + let half = size >> 1; + let middle = first + half; + + let cdf_val = unsafe { *self.cdf.add(middle) }; + + if cdf_val <= u { + first = middle + 1; + size -= half + 1; + } else { + size = half; + } + } + + (first - 1).clamp(0, self.n as usize - 1) + } + pub fn sample(&self, u: Float) -> (Float, Float, usize) { // Find offset via binary search on CDF let offset = self.find_interval(u); @@ -766,12 +787,12 @@ impl DevicePiecewiseConstant2D { // } pub fn integral(&self) -> f32 { - self.p_marginal.integral() + self.marginal.integral() } pub fn sample(&self, u: Point2f) -> (Point2f, f32, Point2i) { - let (d1, pdf1, off_y) = self.p_marginal.sample(u.y()); - let (d0, pdf0, off_x) = self.p_conditional_v[off_y].sample(u.x()); + let (d1, pdf1, off_y) = self.marginal.sample(u.y()); + let (d0, pdf0, off_x) = (unsafe { &*self.conditional.add(off_y) }).sample(u.x()); let pdf = pdf0 * pdf1; let offset = Point2i::new(off_x as i32, off_y as i32); (Point2f::new(d0, d1), pdf, offset) @@ -779,13 +800,13 @@ impl DevicePiecewiseConstant2D { pub fn pdf(&self, p: Point2f) -> Float { // Find which row - let delta_v = 1.0 / self.n_v as Float; + // 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 conditional = unsafe { &*self.conditional.add(v_offset) }; // Find which column - let delta_u = 1.0 / self.n_u as Float; + // 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) }; @@ -1279,7 +1300,7 @@ impl PiecewiseLinear2D { #[inline(always)] fn get_param_value(&self, dim: usize, idx: usize) -> Float { // Safety: Bounds checking against param_size ensures this is valid - unsafe { *self.param_values[dim].0.add(idx) } + unsafe { *self.param_values[dim].add(idx) } } fn get_slice_info(&self, params: [Float; N]) -> (u32, [(Float, Float); N]) { @@ -1341,7 +1362,7 @@ impl PiecewiseLinear2D { current_mask >>= 1; } let idx = (i0 + offset) as usize; - let val = unsafe { *data.0.add(idx) }; + let val = unsafe { *data.add(idx) }; result += weight * val; } result diff --git a/src/core/aggregates.rs b/src/core/aggregates.rs index 116742a..d9558a9 100644 --- a/src/core/aggregates.rs +++ b/src/core/aggregates.rs @@ -1,11 +1,10 @@ -use crate::Float; -use crate::core::geometry::{Bounds3f, Point3f, Ray, Vector3f}; 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::{find_interval, partition_slice}; use rayon::prelude::*; +use shared::Float; +use shared::core::geometry::{Bounds3f, Point3f, Ray, Vector3f}; +use shared::core::shape::ShapeIntersection; +use shared::utils::math::encode_morton_3; +use shared::utils::{find_interval, partition_slice}; use std::cmp::Ordering; use std::sync::Arc; use std::sync::atomic::{AtomicUsize, Ordering as AtomicOrdering}; diff --git a/src/core/camera.rs b/src/core/camera.rs index 8d6bd85..24dd703 100644 --- a/src/core/camera.rs +++ b/src/core/camera.rs @@ -88,6 +88,7 @@ pub trait CameraFactory { medium: Medium, film: Arc, loc: &FileLoc, + arena: &mut Arena, ) -> Result; } diff --git a/src/core/film.rs b/src/core/film.rs index c923473..5ccc8cb 100644 --- a/src/core/film.rs +++ b/src/core/film.rs @@ -1,4 +1,10 @@ -use shared::film::{Film, FilmBase, GBufferFilm, PixelSensor, RGBFilm, SpectralFilm}; +use shared::Float; +use shared::core::spectrum::Spectrum; +use shared::film::{Film, FilmBase, GBufferFilm, PixelSensor, PixelSensor, RGBFilm, SpectralFilm}; +use shared::spectra::{PiecewiseLinearSpectrum, RGBColorSpace}; + +use crate::spectra::DenselySampledSpectrumBuffer; +use crate::utils::{FileLoc, ParameterDictionary}; const N_SWATCH_REFLECTANCES: usize = 24; const SWATCH_REFLECTANCES: Lazy<[Spectrum; N_SWATCH_REFLECTANCES]> = Lazy::new(|| { @@ -10,7 +16,7 @@ const SWATCH_REFLECTANCES: Lazy<[Spectrum; N_SWATCH_REFLECTANCES]> = Lazy::new(| }); pub trait PixelSensorHost { - pub fn get_swatches() -> &[Spectrum; N_SWATCH_REFLECTANCES] { + pub fn get_swatches() -> &'static [Spectrum; N_SWATCH_REFLECTANCES] { &*SWATCH_REFLECTANCES } @@ -29,9 +35,9 @@ pub trait PixelSensorHost { let imaging_ratio = exposure_time * iso / 100.; let d_illum = if white_balance_temp == 0. { - generate_cie_d(6500.) + DenselySampledSpectrumBuffer::generate_cie_d(6500.) } else { - generate_cie_d(white_balance_temp) + DenselySampledSpectrumBuffer::generate_cie_d(white_balance_temp) }; let sensor_illum: Option> = if white_balance_temp != 0. { @@ -73,8 +79,104 @@ pub trait PixelSensorHost { .map_err(|e| e.to_string()); } } + + fn new( + r: Spectrum, + g: Spectrum, + b: Spectrum, + output_colorspace: RGBColorSpace, + sensor_illum: &Spectrum, + imaging_ratio: Float, + spectra: *const StandardSpectra, + swatches: &[Spectrum; 24], + ) -> Self { + // As seen in usages of this constructos, sensor_illum can be null + // Going with the colorspace's own illuminant, but this might not be the right choice + // TODO: Test this + let illum: &Spectrum = match sensor_illum { + Some(arc_illum) => &**arc_illum, + None => &output_colorspace.illuminant, + }; + + let r_bar = DenselySampledSpectrum::from_spectrum(&r); + let g_bar = DenselySampledSpectrum::from_spectrum(&g); + let b_bar = DenselySampledSpectrum::from_spectrum(&b); + let mut rgb_camera = [[0.; 3]; Self::N_SWATCH_REFLECTANCES]; + + let swatches = Self::get_swatches(); + + for i in 0..Self::N_SWATCH_REFLECTANCES { + let rgb = Self::project_reflectance::( + &swatches[i], + illum, + &Spectrum::Dense(r_bar.clone()), + &Spectrum::Dense(g_bar.clone()), + &Spectrum::Dense(b_bar.clone()), + ); + for c in 0..3 { + rgb_camera[i][c] = rgb[c]; + } + } + + let mut xyz_output = [[0.; 3]; Self::N_SWATCH_REFLECTANCES]; + let sensor_white_g = illum.inner_product(&Spectrum::Dense(g_bar.clone())); + let sensor_white_y = illum.inner_product(spectra.y); + for i in 0..Self::N_SWATCH_REFLECTANCES { + let s = swatches[i].clone(); + let xyz = Self::project_reflectance::( + &s, + &Spectrum::Dense(output_colorspace.illuminant), + spectra.x, + spectra.y, + spectra.z, + ) * (sensor_white_y / sensor_white_g); + for c in 0..3 { + xyz_output[i][c] = xyz[c]; + } + } + + let xyz_from_sensor_rgb = linear_least_squares(rgb_camera, xyz_output)?; + + Ok(Self { + xyz_from_sensor_rgb, + r_bar, + g_bar, + b_bar, + imaging_ratio, + }) + } + + fn new_with_white_balance( + output_colorspace: &RGBColorSpace, + sensor_illum: &Spectrum, + imaging_ratio: Float, + spectra: *const StandardSpectra, + ) -> Self { + let r_bar = DenselySampledSpectrumBuffer::from_spectrum(spectra.x); + let g_bar = DenselySampledSpectrumBuffer::from_spectrum(spectra.y); + let b_bar = DenselySampledSpectrumBuffer::from_spectrum(spectra.z); + let xyz_from_sensor_rgb: SquareMatrix; + + if let Some(illum) = sensor_illum { + let source_white = illum.to_xyz(spectra).xy(); + let target_white = output_colorspace.w; + xyz_from_sensor_rgb = white_balance(source_white, target_white); + } else { + xyz_from_sensor_rgb = SquareMatrix::::default(); + } + + Self { + xyz_from_sensor_rgb, + r_bar, + g_bar, + b_bar, + imaging_ratio, + } + } } +impl PixelSensorHost for PixelSensor {} + struct SpectralFilmStorage { pixels: Array2D, bucket_sums: Vec, @@ -164,6 +266,39 @@ impl SpectralFilmHost { } } +impl GBufferFilmHost { + pub fn new( + base: &FilmBase, + output_from_render: &AnimatedTransform, + apply_inverse: bool, + colorspace: &RGBColorSpace, + max_component_value: Float, + write_fp16: bool, + ) -> Self { + assert!(!base.pixel_bounds.is_empty()); + let sensor_ptr = base.sensor; + if sensor_ptr.is_null() { + panic!("Film must have a sensor"); + } + let sensor = unsafe { &*sensor_ptr }; + let output_rgbf_from_sensor_rgb = colorspace.rgb_from_xyz * sensor.xyz_from_sensor_rgb; + let filter_integral = base.filter.integral(); + let pixels = Array2D::new(base.pixel_bounds); + + Self { + base: base.clone(), + output_from_render: output_from_render.clone(), + apply_inverse, + pixels, + colorspace: colorspace.clone(), + max_component_value, + write_fp16, + filter_integral, + output_rgbf_from_sensor_rgb, + } + } +} + pub trait FilmBaseHost { fn create( params: &ParameterDictionary, diff --git a/src/core/filter.rs b/src/core/filter.rs index 7b4b131..3a42ce6 100644 --- a/src/core/filter.rs +++ b/src/core/filter.rs @@ -1,3 +1,5 @@ +use shared::core::filter::FilterSampler; +use shared::core::geometry::Point2f; use shared::filter::Filter; use shared::filters::*; @@ -46,3 +48,37 @@ impl FilterFactory for Filter { } } } + +pub trait CreateFilterSampler { + fn new(radius: Vector2f, func: F) -> Self + where + F: Fn(Point2f) -> Float; +} + +impl CreateFilterSampler for FilterSampler { + fn new(radius: Vector2f, func: F) -> Self + where + F: Fn(Point2f) -> Float, + { + let domain = Bounds2f::from_points( + Point2f::new(-radius.x(), -radius.y()), + Point2f::new(radius.x(), radius.y()), + ); + + let nx = (32.0 * radius.x()) as usize; + let ny = (32.0 * radius.y()) as usize; + + let mut f = Array2D::new_with_dims(nx, ny); + for y in 0..f.y_size() { + for x in 0..f.x_size() { + let p = domain.lerp(Point2f::new( + (x as Float + 0.5) / f.x_size() as Float, + (y as Float + 0.5) / f.y_size() as Float, + )); + f[(x as i32, y as i32)] = func(p); + } + } + let distrib = DevicePiecewiseConstant2D::new_with_bounds(&f, domain); + Self { domain, f, distrib } + } +} diff --git a/src/core/image/io.rs b/src/core/image/io.rs index 921374c..482fec3 100644 --- a/src/core/image/io.rs +++ b/src/core/image/io.rs @@ -1,12 +1,11 @@ use super::{Image, ImageAndMetadata}; +use crate::core::image::PixelStorage; use crate::utils::error::ImageError; use anyhow::{Context, Result, bail}; use exr::prelude::{read_first_rgba_layer_from_file, write_rgba_file}; -use image_rs::ImageReader; -use image_rs::{DynamicImage, ImageBuffer, Rgb, Rgba}; +use image_rs::{DynamicImage, ImageReader}; use shared::Float; use shared::core::color::{ColorEncoding, LINEAR, SRGB}; -use shared::core::image::DeviceImage; use std::fs::File; use std::io::{BufRead, BufReader, BufWriter, Read, Write}; use std::path::Path; @@ -182,12 +181,12 @@ impl ImageIO for Image { fn to_u8_buffer(&self) -> Vec { match &self.pixels { - PixelData::U8(data) => data.clone(), - PixelData::F16(data) => data + PixelStorage::U8(data) => data.clone(), + PixelStorage::F16(data) => data .iter() .map(|v| (v.to_f32().clamp(0.0, 1.0) * 255.0 + 0.5) as u8) .collect(), - PixelData::F32(data) => data + PixelStorage::F32(data) => data .iter() .map(|v| (v.clamp(0.0, 1.0) * 255.0 + 0.5) as u8) .collect(), @@ -205,41 +204,27 @@ fn read_generic(path: &Path, encoding: Option) -> Result 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) => Image { - format: PixelFormat::F32, - resolution: res, - channel_names: vec!["R".into(), "G".into(), "B".into(), "A".into()], - encoding: LINEAR, - pixels: PixelData::F32(buf.into_raw()), - }, + DynamicImage::ImageRgb32F(buf) => Image::from_f32(buf.into_raw(), res, rgb_names()), + DynamicImage::ImageRgba32F(buf) => Image::from_f32(buf.into_raw(), res, rgba_names()), _ => { // Default to RGB8 for everything else + let enc = encoding.unwrap_or(ColorEncoding::sRGB); if dyn_img.color().has_alpha() { let buf = dyn_img.to_rgba8(); - Image { - format: PixelFormat::U8, - resolution: res, - channel_names: vec!["R".into(), "G".into(), "B".into(), "A".into()], - encoding: encoding.unwrap_or(SRGB), - pixels: PixelData::U8(buf.into_raw()), - } + Image::from_u8(buf.into_raw(), res, rgba_names(), enc) } else { let buf = dyn_img.to_rgb8(); - Image { - format: PixelFormat::U8, - resolution: res, - channel_names: vec!["R".into(), "G".into(), "B".into()], - encoding: encoding.unwrap_or(SRGB), - pixels: PixelData::U8(buf.into_raw()), - } + Image::from_u8(buf.into_raw(), res, rgb_names(), enc) } } }; @@ -275,7 +260,7 @@ fn read_exr(path: &Path) -> Result { resolution: Point2i::new(w, h), channel_names: vec!["R".into(), "G".into(), "B".into(), "A".into()], encoding: LINEAR, - pixels: PixelData::F32(image.layer_data.channel_data.pixels), + pixels: PixelStorage::F32(image.layer_data.channel_data.pixels), }; let metadata = ImageMetadata::default(); @@ -360,7 +345,7 @@ fn read_pfm(path: &Path) -> Result { resolution: Point2i::new(w, h), channel_names: names, encoding: LINEAR, - pixels: PixelData::F32(pixels), + pixels: PixelStorage::F32(pixels), }; let metadata = ImageMetadata::default(); diff --git a/src/core/image/metadata.rs b/src/core/image/metadata.rs index e8aef14..f87bee0 100644 --- a/src/core/image/metadata.rs +++ b/src/core/image/metadata.rs @@ -1,11 +1,11 @@ -use crate::core::geometry::{Bounds2i, Point2i}; -use crate::core::pbrt::Float; -use crate::spectra::colorspace::RGBColorSpace; -use crate::utils::math::SquareMatrix; -use smallvec::SmallVec; +use shared::Float; +use shared::core::geometry::{Bounds2i, Point2i}; +use shared::spectra::RGBColorSpace; +use shared::utils::math::SquareMatrix; use std::collections::HashMap; -use std::ops::{Deref, DerefMut}; +// use std::ops::{Deref, DerefMut}; + #[derive(Debug, Clone, Default)] pub struct ImageChannelDesc { pub offset: Vec, diff --git a/src/core/image/mod.rs b/src/core/image/mod.rs index b7d4e22..24efaf6 100644 --- a/src/core/image/mod.rs +++ b/src/core/image/mod.rs @@ -1,6 +1,7 @@ +use half::f16; use shared::core::geometry::Point2i; use shared::core::image::{DeviceImage, ImageAccess, ImageBase, PixelFormat, WrapMode}; -use shared::utils::math::f16_to_f32; +use smallvec::smallvec; use std::ops::Deref; pub mod io; @@ -11,16 +12,6 @@ pub mod pixel; pub use io::ImageIO; pub use metadata::*; -impl std::fmt::Display for PixelFormat { - fn fmt(&self, f: &mut std::fmt::Formatter<'_>) -> std::fmt::Result { - match self { - PixelFormat::U8 => write!(f, "U256"), - PixelFormat::F16 => write!(f, "Half"), - PixelFormat::F32 => write!(f, "Float"), - } - } -} - #[derive(Clone, Debug, Default)] pub struct ImageChannelValues(pub SmallVec<[Float; 4]>); diff --git a/src/core/image/ops.rs b/src/core/image/ops.rs index 11b63cd..ca50738 100644 --- a/src/core/image/ops.rs +++ b/src/core/image/ops.rs @@ -1,5 +1,5 @@ -// use rayon::prelude::*; use super::Image; +use crate::core::image::PixelStorage; use rayon::prelude::*; use shared::Float; use shared::core::geometry::{Bounds2i, Point2i}; @@ -19,9 +19,9 @@ impl Image { let nc = self.n_channels(); match &mut self.pixels { - PixelData::U8(d) => flip_y_kernel(d, res, nc), - PixelData::F16(d) => flip_y_kernel(d, res, nc), - PixelData::F32(d) => flip_y_kernel(d, res, nc), + PixelStorage::U8(d) => flip_y_kernel(d, res, nc), + PixelStorage::F16(d) => flip_y_kernel(d, res, nc), + PixelStorage::F32(d) => flip_y_kernel(d, res, nc), } } @@ -39,13 +39,13 @@ impl Image { ); match (&self.pixels, &mut new_image.pixels) { - (PixelData::U8(src), PixelData::U8(dst)) => { + (PixelStorage::U8(src), PixelStorage::U8(dst)) => { crop_kernel(src, dst, self.resolution, bounds, self.n_channels()) } - (PixelData::F16(src), PixelData::F16(dst)) => { + (PixelStorage::F16(src), PixelStorage::F16(dst)) => { crop_kernel(src, dst, self.resolution, bounds, self.n_channels()) } - (PixelData::F32(src), PixelData::F32(dst)) => { + (PixelStorage::F32(src), PixelStorage::F32(dst)) => { crop_kernel(src, dst, self.resolution, bounds, self.n_channels()) } _ => panic!("Format mismatch in crop"), @@ -56,9 +56,9 @@ impl Image { pub fn copy_rect_out(&self, extent: Bounds2i, buf: &mut [Float], wrap: WrapMode2D) { match &self.pixels { - PixelData::U8(d) => copy_rect_out_kernel(d, self, extent, buf, wrap), - PixelData::F16(d) => copy_rect_out_kernel(d, self, extent, buf, wrap), - PixelData::F32(d) => copy_rect_out_kernel(d, self, extent, buf, wrap), + PixelStorage::U8(d) => copy_rect_out_kernel(d, self, extent, buf, wrap), + PixelStorage::F16(d) => copy_rect_out_kernel(d, self, extent, buf, wrap), + PixelStorage::F32(d) => copy_rect_out_kernel(d, self, extent, buf, wrap), } } @@ -68,13 +68,13 @@ impl Image { let encoding = self.encoding; match &mut self.pixels { - PixelData::U8(d) => { + PixelStorage::U8(d) => { copy_rect_in_kernel(d, resolution, n_channels, encoding, extent, buf) } - PixelData::F16(d) => { + PixelStorage::F16(d) => { copy_rect_in_kernel(d, resolution, n_channels, encoding, extent, buf) } - PixelData::F32(d) => { + PixelStorage::F32(d) => { copy_rect_in_kernel(d, resolution, n_channels, encoding, extent, buf) } } @@ -154,9 +154,9 @@ impl Image { ); match &mut next.pixels { - PixelData::U8(d) => downsample_kernel(d, new_res, prev, internal_wrap), - PixelData::F16(d) => downsample_kernel(d, new_res, prev, internal_wrap), - PixelData::F32(d) => downsample_kernel(d, new_res, prev, internal_wrap), + PixelStorage::U8(d) => downsample_kernel(d, new_res, prev, internal_wrap), + PixelStorage::F16(d) => downsample_kernel(d, new_res, prev, internal_wrap), + PixelStorage::F32(d) => downsample_kernel(d, new_res, prev, internal_wrap), } levels.push(next); } diff --git a/src/core/light.rs b/src/core/light.rs index 4ec1c76..2e146e3 100644 --- a/src/core/light.rs +++ b/src/core/light.rs @@ -1,16 +1,16 @@ -use crate::core::spectrum::{SPECTRUM_CACHE, spectrum_to_photometric}; +use crate::core::spectrum::SPECTRUM_CACHE; use crate::core::texture::FloatTexture; use crate::lights::*; use crate::utils::containers::InternCache; -use crate::utils::{Arena, FileLoc, ParameterDictionary, Upload, resolve_filename}; +use crate::utils::{Arena, FileLoc, ParameterDictionary}; +use log::error; 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}; +use shared::utils::Transform; pub fn lookup_spectrum(s: &Spectrum) -> DenselySampledSpectrum { let cache = SPECTRUM_CACHE.get_or_init(InternCache::new); @@ -120,7 +120,7 @@ impl LightFactory for Light { alpha_tex, colorspace, )?, - "infinite" => infinite::create( + "infinite" => crate::lights::infinite::create( arena, render_from_light, medium, diff --git a/src/core/material.rs b/src/core/material.rs index 737d2f9..7c143cd 100644 --- a/src/core/material.rs +++ b/src/core/material.rs @@ -1,6 +1,7 @@ +use crate::Arena; use crate::core::image::Image; +use crate::utils::TextureParameterDictionary; use crate::utils::error::FileLoc; -use crate::utils::parameters::ParameterDictionary; use shared::core::material::Material; use shared::materials::*; use std::collections::HashMap; @@ -11,6 +12,7 @@ pub trait CreateMaterial: Sized { normal_map: Option>, named_materials: &HashMap, loc: &FileLoc, + arena: &mut Arena, ) -> Result; } @@ -38,6 +40,7 @@ pub trait MaterialFactory { normal_map: Ptr, named_materials: HashMap, loc: &FileLoc, + arena: &mut Arena, ) -> Result; } @@ -48,7 +51,8 @@ impl MaterialFactory for Material { normal_map: Option>, named_materials: &HashMap, loc: &FileLoc, - ) -> Result { + arena: &mut Arena, + ) -> Result { make_material_factory!( name, params, normal_map, named_materials, loc; diff --git a/src/core/mod.rs b/src/core/mod.rs index 7e673a4..b64cb31 100644 --- a/src/core/mod.rs +++ b/src/core/mod.rs @@ -7,6 +7,7 @@ pub mod filter; pub mod image; pub mod light; pub mod material; +pub mod medium; pub mod sampler; pub mod sampler; pub mod scene; diff --git a/src/core/sampler.rs b/src/core/sampler.rs index 1c07da3..a6816f8 100644 --- a/src/core/sampler.rs +++ b/src/core/sampler.rs @@ -1,12 +1,24 @@ use shared::core::sampler::Sampler; +use crate::Arena; + +pub trait CreateSampler { + fn create( + params: &ParameterDictionary, + full_res: Point2i, + loc: &FileLoc, + arena: &mut Arena, + ) -> Result; +} + pub trait SamplerFactory { fn create( name: &str, params: &ParameterDictionary, full_res: Point2i, loc: &FileLoc, - ) -> Result; + arena: &mut Arena, + ) -> Result; } impl SamplerFactory for Sampler { @@ -15,30 +27,31 @@ impl SamplerFactory for Sampler { params: &ParameterDictionary, full_res: Point2i, loc: &FileLoc, - ) -> Result { + arena: &mut Arena, + ) -> Result { match name { "zsobol" => { - let sampler = ZSobolSampler::create(params, full_res, loc)?; + let sampler = ZSobolSampler::create(params, full_res, loc, arena)?; Ok(Sampler::ZSobol(sampler)) } "paddedsobol" => { - let sampler = PaddedSobolSampler::create(params, full_res, loc)?; + let sampler = PaddedSobolSampler::create(params, full_res, loc, arena)?; Ok(Sampler::PaddedSobol(sampler)) } "halton" => { - let sampler = HaltonSampler::create(params, full_res, loc)?; + let sampler = HaltonSampler::create(params, full_res, loc, arena)?; Ok(Sampler::Halton(sampler)) } "sobol" => { - let sampler = SobolSampler::create(params, full_res, loc)?; + let sampler = SobolSampler::create(params, full_res, loc, arena)?; Ok(Sampler::Sobol(sampler)) } "Independent" => { - let sampler = IndependentSampler::create(params, full_res, loc)?; + let sampler = IndependentSampler::create(params, full_res, loc, arena)?; Ok(Sampler::Independent(sampler)) } "stratified" => { - let sampler = StratifiedSampler::create(params, full_res, loc)?; + let sampler = StratifiedSampler::create(params, full_res, loc, arena)?; Ok(Sampler::Stratified(sampler)) } _ => Err(format!("Film type '{}' unknown at {}", name, loc)), diff --git a/src/core/scene.rs b/src/core/scene.rs index bb75ec5..4ac7d81 100644 --- a/src/core/scene.rs +++ b/src/core/scene.rs @@ -12,27 +12,27 @@ use image_rs::Primitive; use parking_lot::Mutex; use shared::Float; use shared::core::camera::{Camera, CameraTransform}; -use shared::core::color::LINEAR; +// use shared::core::color::LINEAR; use shared::core::film::{Film, FilmTrait}; use shared::core::filter::Filter; -use shared::core::geometry::{Point3f, Vector3f}; +use shared::core::geometry::Vector3f; use shared::core::lights::Light; use shared::core::material::Material; use shared::core::medium::{Medium, MediumInterface}; use shared::core::options::RenderingCoordinateSystem; -use shared::core::primitive::{Primitive, PrimitiveTrait}; +use shared::core::primitive::PrimitiveTrait; use shared::core::sampler::Sampler; -use shared::core::spectrum::{Spectrum, SpectrumType}; +use shared::core::spectrum::SpectrumType; use shared::core::texture::{FloatTexture, SpectrumTexture}; use shared::images::Image; use shared::spectra::RGBColorSpace; use shared::utils::error::FileLoc; -use shared::utils::math::SquareMatrix; +// use shared::utils::math::SquareMatrix; use shared::utils::transform::{AnimatedTransform, Transform, look_at}; use std::collections::{HashMap, HashSet}; use std::ops::{Index as IndexTrait, IndexMut as IndexMutTrait}; use std::sync::Arc; -use std::sync::atomic::{AtomicI32, Ordering}; +// use std::sync::atomic::{AtomicI32, Ordering}; #[derive(Clone, Debug)] pub enum MaterialRef { @@ -1000,6 +1000,7 @@ struct GraphicsState { pub current_inside_medium: String, pub current_outside_medium: String, pub current_material_name: String, + pub current_material_index: Option, pub area_light_name: String, pub area_light_params: ParsedParameterVector, pub area_light_loc: FileLoc, @@ -1064,6 +1065,7 @@ impl BasicSceneBuilder { named_coordinate_systems: HashMap::new(), active_instance_definition: None, shapes: Vec::new(), + animated_shapes: Vec::new(), named_material_names: HashSet::new(), medium_names: HashSet::new(), @@ -1526,8 +1528,9 @@ impl ParserTarget for BasicSceneBuilder { } } - fn material(&mut self, _name: &str, _params: &ParsedParameterVector, _loc: FileLoc) { - todo!() + fn material(&mut self, _name: &str, _params: &ParsedParameterVector, loc: FileLoc) { + self.verify_world("material", loc); + self.graphics_state.current_material_name = self.scene.add_material(name, material) } fn make_named_material(&mut self, _name: &str, _params: &ParsedParameterVector, _loc: FileLoc) { todo!() diff --git a/src/core/shape.rs b/src/core/shape.rs index e81ffe2..5a0e933 100644 --- a/src/core/shape.rs +++ b/src/core/shape.rs @@ -1,16 +1,15 @@ use crate::core::texture::FloatTexture; -use crate::shapes::TriQuadMesh; +use crate::shapes::{BilinearPatchMeshHost, TriQuadMesh, TriangleMeshHost}; use crate::utils::{Arena, FileLoc, ParameterDictionary, resolve_filename}; use shared::core::options::get_options; use shared::core::shape::*; use shared::shapes::*; -use shared::spectra::*; +// use shared::spectra::*; use shared::utils::Transform; -use shared::utils::mesh::{BilinearPatchMesh, TriangleMesh}; use std::sync::{Arc, Mutex}; pub static ALL_TRIANGLE_MESHES: Mutex>> = Mutex::new(Vec::new()); -pub static ALL_TRIANGLE_MESHES: Mutex>> = Mutex::new(Vec::new()); +pub static ALL_BILINEAR_MESHES: Mutex>> = Mutex::new(Vec::new()); pub trait CreateShape { fn create( diff --git a/src/core/spectrum.rs b/src/core/spectrum.rs index 3114e40..75401d3 100644 --- a/src/core/spectrum.rs +++ b/src/core/spectrum.rs @@ -2,7 +2,6 @@ use crate::core::light::LightBaseTrait; use crate::spectra::cie_y; use crate::utils::containers::InternCache; use shared::Float; -use shared::core::light::LightBase; use shared::core::spectrum::Spectrum; pub static SPECTRUM_CACHE: Lazy>> = diff --git a/src/core/texture.rs b/src/core/texture.rs index 19e37cb..b88a1d7 100644 --- a/src/core/texture.rs +++ b/src/core/texture.rs @@ -1,7 +1,7 @@ use crate::textures::*; -use crate::utils::FileLoc; use crate::utils::mipmap::MIPMapFilterOptions; -use crate::utils::{FileLoc, TextureParameterDictionary}; +use crate::utils::{Arena, FileLoc, TextureParameterDictionary}; +use enum_dispatch::enum_dispatch; use shared::textures::*; use shared::utils::Transform; use std::sync::{Arc, Mutex, OnceLock}; @@ -29,19 +29,25 @@ impl FloatTexture { render_from_texture: &Transform, params: &TextureParameterDictionary, loc: &FileLoc, + arena: &mut Arena, ) -> Result { match name { "constant" => { let tex = FloatConstantTexture::create(render_from_texture, params, loc); Ok(FloatTexture::Constant(tex)) } - "scale" => Ok(FloatScaledTexture::create(render_from_texture, params, loc)), + "scale" => Ok(FloatScaledTexture::create( + render_from_texture, + params, + loc, + arena, + )), "mix" => { - let tex = FloatMixTexture::create(render_from_texture, params, loc); + let tex = FloatMixTexture::create(render_from_texture, params, loc, arena); Ok(FloatTexture::Mix(tex)) } "directionmix" => { - let tex = FloatDirectionMixTexture::create(render_from_texture, params, loc); + let tex = FloatDirectionMixTexture::create(render_from_texture, params, loc, arena); Ok(FloatTexture::DirectionMix(tex)) } "bilerp" => { @@ -50,7 +56,7 @@ impl FloatTexture { } "imagemap" => { let tex = FloatImageTexture::create(render_from_texture, params, loc); - Ok(loatTexture::Image(tex)) + Ok(FloatTexture::Image(tex)) } "checkerboard" => { let tex = FloatCheckerboardTexture::create(render_from_texture, params, loc); @@ -143,10 +149,11 @@ pub static TEXTURE_CACHE: OnceLock>>> = OnceL pub fn get_texture_cache() -> &'static Mutex>> { TEXTURE_CACHE.get_or_init(|| Mutex::new(HashMap::new())) } + #[derive(Debug, Hash, PartialEq, Eq, Clone)] -struct TexInfo { - filename: String, - filter_options: MIPMapFilterOptions, - wrap_mode: WrapMode, - encoding: ColorEncoding, +pub struct TexInfo { + pub filename: String, + pub filter_options: MIPMapFilterOptions, + pub wrap_mode: WrapMode, + pub encoding: ColorEncoding, } diff --git a/src/integrators/mod.rs b/src/integrators/mod.rs index 0c9797f..7749e02 100644 --- a/src/integrators/mod.rs +++ b/src/integrators/mod.rs @@ -1,9 +1,7 @@ mod pipeline; -use pipeline::*; - use shared::core::bssrdf::{BSSRDFTrait, SubsurfaceInteraction}; -use shared::core::bxdf::{BSDF, BxDFFlags, BxDFReflTransFlags, BxDFTrait, FArgs, TransportMode}; +use shared::core::bxdf::{BSDF, BxDFFlags, BxDFTrait, FArgs, TransportMode}; use shared::core::camera::Camera; use shared::core::film::VisibleSurface; use shared::core::geometry::{Bounds2i, Point2f, Point2i, Point3fi, Ray, Vector3f, VectorLike}; @@ -15,7 +13,7 @@ use shared::core::medium::{MediumTrait, PhaseFunctionTrait}; use shared::core::options::get_options; use shared::core::pbrt::{Float, SHADOW_EPSILON}; use shared::core::primitive::{Primitive, PrimitiveTrait}; -use shared::core::sampler::{CameraSample, Sampler, SamplerTrait}; +use shared::core::sampler::{Sampler, SamplerTrait}; use shared::lights::sampler::LightSamplerTrait; use shared::lights::sampler::{LightSampler, UniformLightSampler}; use shared::shapes::ShapeIntersection; @@ -30,6 +28,8 @@ use shared::utils::sampling::{ use std::sync::Arc; +use crate::Arena; + #[derive(Clone, Debug)] pub struct IntegratorBase { pub aggregate: Arc, @@ -70,7 +70,13 @@ pub trait IntegratorTrait { } pub trait RayIntegratorTrait { - fn evaluate_pixel_sample(&self, p_pixel: Point2i, sample_ind: usize, sampler: &mut Sampler); + fn evaluate_pixel_sample( + &self, + p_pixel: Point2i, + sample_ind: usize, + sampler: &mut Sampler, + arena: &mut Arena, + ); fn li( &self, @@ -78,6 +84,7 @@ pub trait RayIntegratorTrait { lambda: &SampledWavelengths, sampler: &mut Sampler, visible_surface: bool, + arena: &mut Arena, ) -> (SampledSpectrum, Option); } @@ -201,8 +208,21 @@ impl SimplePathIntegrator { } impl RayIntegratorTrait for SimplePathIntegrator { - fn evaluate_pixel_sample(&self, p_pixel: Point2i, sample_ind: usize, sampler: &mut Sampler) { - pipeline::evaluate_pixel_sample(self, self.camera.as_ref(), sampler, p_pixel, sample_ind); + fn evaluate_pixel_sample( + &self, + p_pixel: Point2i, + sample_ind: usize, + sampler: &mut Sampler, + arena: &mut Arena, + ) { + pipeline::evaluate_pixel_sample( + self, + self.camera.as_ref(), + sampler, + p_pixel, + sample_ind, + arena, + ); } fn li( @@ -211,6 +231,7 @@ impl RayIntegratorTrait for SimplePathIntegrator { lambda: &SampledWavelengths, sampler: &mut Sampler, _visible_surface: bool, + arena: &mut Arena, ) -> (SampledSpectrum, Option) { let mut l = SampledSpectrum::new(0.0); let mut beta = SampledSpectrum::new(1.0); @@ -333,7 +354,7 @@ impl RayIntegratorTrait for PathIntegrator { p_pixel: Point2i, sample_ind: usize, sampler: &mut Sampler, - scratch: &Bump, + arena: &mut Arena, ) { pipeline::evaluate_pixel_sample( self, @@ -341,7 +362,7 @@ impl RayIntegratorTrait for PathIntegrator { sampler, p_pixel, sample_ind, - scratch, + arena, ); } @@ -350,8 +371,8 @@ impl RayIntegratorTrait for PathIntegrator { mut ray: Ray, lambda: &SampledWavelengths, sampler: &mut Sampler, - _scratch: &Bump, visible_surface: bool, + _arena: &mut Arena, ) -> (SampledSpectrum, Option) { let mut l = SampledSpectrum::new(0.0); let mut beta = SampledSpectrum::new(1.0); @@ -508,7 +529,7 @@ impl RayIntegratorTrait for SimpleVolPathIntegrator { p_pixel: Point2i, sample_ind: usize, sampler: &mut Sampler, - scratch: &Bump, + arena: &mut Arena, ) { pipeline::evaluate_pixel_sample( self, @@ -516,7 +537,7 @@ impl RayIntegratorTrait for SimpleVolPathIntegrator { sampler, p_pixel, sample_ind, - scratch, + arena, ); } @@ -525,8 +546,8 @@ impl RayIntegratorTrait for SimpleVolPathIntegrator { mut ray: Ray, lambda: &SampledWavelengths, sampler: &mut Sampler, - _scratch: &Bump, _visible_surface: bool, + _arena: &mut Arena, ) -> (SampledSpectrum, Option) { let mut l = SampledSpectrum::new(0.); let mut beta = 1.; @@ -786,7 +807,7 @@ impl RayIntegratorTrait for VolPathIntegrator { p_pixel: Point2i, sample_ind: usize, sampler: &mut Sampler, - scratch: &Bump, + arena: &mut Arena, ) { pipeline::evaluate_pixel_sample( self, @@ -803,8 +824,8 @@ impl RayIntegratorTrait for VolPathIntegrator { mut ray: Ray, lambda: &SampledWavelengths, sampler: &mut Sampler, - scratch: &Bump, visible_surface: bool, + arena: &mut Arena, ) -> (SampledSpectrum, Option) { let mut l = SampledSpectrum::new(0.); let mut beta = SampledSpectrum::new(1.); diff --git a/src/integrators/pipeline.rs b/src/integrators/pipeline.rs index 4b6490d..827c430 100644 --- a/src/integrators/pipeline.rs +++ b/src/integrators/pipeline.rs @@ -1,5 +1,6 @@ use crate::core::image::Image; use crate::core::{options::PBRTOptions, sampler::get_camera_sample}; +use crate::spectra::get_spectra_context; use indicatif::{ProgressBar, ProgressStyle}; use std::io::Write; use std::path::Path; @@ -66,6 +67,7 @@ pub fn render( _base: &IntegratorBase, camera: &Camera, sampler_prototype: &Sampler, + arena: &mut Arena, ) where T: RayIntegratorTrait, { @@ -76,7 +78,14 @@ pub fn render( tile_sampler.start_pixel_sample(p_pixel, s_index, None); - evaluate_pixel_sample(integrator, camera, &mut tile_sampler, p_pixel, s_index); + evaluate_pixel_sample( + integrator, + camera, + &mut tile_sampler, + p_pixel, + s_index, + arena, + ); return; } @@ -142,7 +151,14 @@ pub fn render( for p_pixel in tile_bounds { for sample_index in wave_start..wave_end { sampler.start_pixel_sample(*p_pixel, sample_index, None); - evaluate_pixel_sample(integrator, camera, &mut sampler, *p_pixel, sample_index); + evaluate_pixel_sample( + integrator, + camera, + &mut sampler, + *p_pixel, + sample_index, + arena, + ); } } @@ -202,6 +218,7 @@ pub fn evaluate_pixel_sample( sampler: &mut Sampler, pixel: Point2i, _sample_index: usize, + arena: &mut Arena, ) { let mut lu = sampler.get1d(); if get_options().disable_wavelength_jitter { @@ -221,10 +238,11 @@ pub fn evaluate_pixel_sample( let initialize_visible_surface = film.uses_visible_surface(); let (mut l, visible_surface) = - integrator.li(camera_ray.ray, &lambda, sampler, initialize_visible_surface); + integrator.li(camera_ray.ray, &lambda, sampler, initialize_visible_surface, arena); l *= camera_ray.weight; - if l.has_nans() || l.y(&lambda).is_infinite() { + let std_spectra = get_spectra_context(); + if l.has_nans() || l.y(&lambda, std_spectra).is_infinite() { l = SampledSpectrum::new(0.); } diff --git a/src/lib.rs b/src/lib.rs index e2db169..3052ac3 100644 --- a/src/lib.rs +++ b/src/lib.rs @@ -1,5 +1,7 @@ pub mod core; +pub mod filters; pub mod globals; +pub mod gpu; pub mod integrators; pub mod lights; pub mod materials; @@ -8,3 +10,5 @@ pub mod shapes; pub mod spectra; pub mod textures; pub mod utils; + +pub use utils::arena::Arena; diff --git a/src/lights/diffuse.rs b/src/lights/diffuse.rs index a8ce937..7301d61 100644 --- a/src/lights/diffuse.rs +++ b/src/lights/diffuse.rs @@ -1,22 +1,19 @@ +use crate::core::image::{Image, ImageIO}; +use crate::core::light::{CreateLight, lookup_spectrum}; +use crate::core::spectrum::spectrum_to_photometric; +use crate::core::texture::FloatTexture; +use crate::utils::{Arena, FileLoc, ParameterDictionary, Upload, resolve_filename}; +use log::error; use shared::Float; use shared::core::ligh::{Light, LightBase, LightType}; use shared::core::medium::MediumInterface; use shared::core::shape::Shape; use shared::core::spectrum::{Spectrum, SpectrumTrait}; use shared::core::texture::SpectrumType; +use shared::core::texture::{FloatTextureTrait, TextureEvalContext}; use shared::lights::DiffuseAreaLight; use shared::spectra::RGBColorSpace; -use shared::utils::Transform; -use std::sync::Arc; - -use crate::core::image::{Image, ImageIO}; -use crate::core::light::{CreateLight, lookup_spectrum}; -use crate::core::spectrum::spectrum_to_photometric; -use crate::core::texture::{ - FloatTexture, FloatTextureTrait, TextureEvalContext, TextureEvaluator, - UniversalTextureEvaluator, -}; -use crate::utils::{Arena, FileLoc, ParameterDictionary, Upload, resolve_filename}; +use shared::utils::{Ptr, Transform}; pub trait CreateDiffuseLight { fn new( @@ -59,7 +56,7 @@ impl CreateDiffuseLight for DiffuseAreaLight { let base = LightBase::new(light_type, &render_from_light, &medium_interface); - let lemit = LightBase::lookup_spectrum(&le); + let lemit = lookup_spectrum(&le); if let Some(im) = &image { let desc = im .get_channel_desc(&["R", "G", "B"]) diff --git a/src/lights/distant.rs b/src/lights/distant.rs index 1a23469..5117aad 100644 --- a/src/lights/distant.rs +++ b/src/lights/distant.rs @@ -2,13 +2,13 @@ use crate::core::light::{CreateLight, lookup_spectrum}; use crate::core::spectrum::spectrum_to_photometric; use crate::core::texture::FloatTexture; use crate::utils::{Arena, FileLoc, ParameterDictionary}; -use shared::core::geometry::{Vector3f, VectorLike}; +use shared::core::geometry::{Point3f, Vector3f, VectorLike}; use shared::core::light::{Light, LightBase}; use shared::core::medium::{Medium, MediumInterface}; use shared::core::shape::Shape; use shared::lights::DistantLight; use shared::spectra::RGBColorSpace; -use shared::utils::{Ptr, Transform}; +use shared::utils::Transform; pub trait CreateDistantLight { fn new(render_from_light: Transform, le: Spectrum, scale: Float) -> Self; @@ -53,7 +53,7 @@ impl CreateLight for DistantLight { let mut scale = parameters.get_one_float("scale", 1); let from = parameters.get_one_point3f("from", Point3f::new(0., 0., 0.)); - let to = parameters.get_one_point3f("to", Point3f(0., 0., 1.)); + let to = parameters.get_one_point3f("to", Point3f::new(0., 0., 1.)); let w = (from - to).normalize(); let (v1, v2) = w.coordinate_system(); let m: [Float; 16] = [ diff --git a/src/lights/goniometric.rs b/src/lights/goniometric.rs index c0bb752..720cfc5 100644 --- a/src/lights/goniometric.rs +++ b/src/lights/goniometric.rs @@ -1,19 +1,18 @@ -use crate::core::image::{Image, ImageIO, ImageMetadata}; +use crate::core::image::{Image, ImageIO}; 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 log::error; use shared::Float; -use shared::core::image::ImageBase; use shared::core::light::{Light, LightBase, LightType}; use shared::core::medium::{Medium, MediumInterface}; use shared::core::spectrum::Spectrum; use shared::core::texture::SpectrumType; use shared::lights::GoniometricLight; use shared::spectra::RGBColorSpace; -use shared::utils::containers::Array2D; use shared::utils::{Ptr, Transform}; pub trait CreateGoniometricLight { diff --git a/src/lights/infinite.rs b/src/lights/infinite.rs index 1bcec19..47efbc9 100644 --- a/src/lights/infinite.rs +++ b/src/lights/infinite.rs @@ -1,5 +1,6 @@ +use log::error; use shared::Float; -use shared::core::geometry::{Bounds3f, Point2f}; +use shared::core::geometry::Point2f; use shared::core::light::{CreateLight, Light, LightBase, LightType}; use shared::core::medium::MediumInterface; use shared::core::spectrum::Spectrum; diff --git a/src/lights/point.rs b/src/lights/point.rs index 9aaf4c5..89ae8ca 100644 --- a/src/lights/point.rs +++ b/src/lights/point.rs @@ -2,7 +2,6 @@ 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::core::geometry::VectorLike; use shared::core::light::{Light, LightBase, LightType}; use shared::core::medium::{Medium, MediumInterface}; use shared::core::shape::Shape; diff --git a/src/lights/projection.rs b/src/lights/projection.rs index 0b4569b..428e552 100644 --- a/src/lights/projection.rs +++ b/src/lights/projection.rs @@ -1,8 +1,9 @@ -use crate::core::image::{Image, ImageIO, ImageMetadata}; +use crate::core::image::{Image, ImageIO}; use crate::core::light::CreateLight; use crate::core::spectrum::spectrum_to_photometric; use crate::spectra::colorspace::new; -use crate::utils::{Arena, ParameterDictionary, Ptr, Upload, resolve_filename}; +use crate::utils::{Arena, ParameterDictionary, Upload, resolve_filename}; +use log::error; use shared::Float; use shared::core::geometry::{Bounds2f, VectorLike}; use shared::core::image::ImageAccess; diff --git a/src/lights/spot.rs b/src/lights/spot.rs index f944154..e956213 100644 --- a/src/lights/spot.rs +++ b/src/lights/spot.rs @@ -1,18 +1,16 @@ -use crate::core::image::{Image, ImageIO, ImageMetadata}; +// 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::{Arena, ParameterDictionary, Ptr, Upload, resolve_filename}; -use shared::core::geometry::{Bounds2f, Frame, VectorLike}; -use shared::core::image::ImageAccess; +use crate::utils::{Arena, ParameterDictionary}; +use shared::core::geometry::{Frame, VectorLike}; use shared::core::light::{Light, LightBase, LightType}; use shared::core::medium::MediumInterface; use shared::core::spectrum::Spectrum; use shared::core::texture::SpectrumType; -use shared::lights::{ProjectionLight, SpotLight}; +use shared::lights::SpotLight; use shared::spectra::RGBColorSpace; -use shared::utils::math::{radians, square}; -use shared::utils::{Ptr, Transform}; +use shared::utils::Transformk +use shared::utils::math::radians; use shared::{Float, PI}; pub trait CreateSpotLight { @@ -32,7 +30,6 @@ impl CreateSpotLight for SpotLight { medium_interface: MediumInterface, le: Spectrum, scale: shared::Float, - image: Ptr, cos_falloff_start: Float, total_width: Float, ) -> Self { @@ -98,6 +95,7 @@ impl CreateLight for SpotLight { coneangle, coneangle - conedelta, ); + arena.alloc(specific); Ok(Light::Spot(specific)) } } diff --git a/src/materials/coated.rs b/src/materials/coated.rs index 6aceb7f..65979d6 100644 --- a/src/materials/coated.rs +++ b/src/materials/coated.rs @@ -1,11 +1,14 @@ use crate::core::image::Image; use crate::core::material::CreateMaterial; -use crate::utils::{Arena, ParameterDictionary}; +use crate::core::texture::SpectrumTexture; +use crate::utils::{Arena, FileLoc, TextureParameterDictionary, Upload}; +use shared::core::material::Material; use shared::core::spectrum::Spectrum; use shared::core::texture::SpectrumType; use shared::materials::coated::*; use shared::spectra::ConstantSpectrum; use shared::textures::SpectrumConstantTexture; +use std::collections::HashMap; impl CreateMaterial for CoatedDiffuseMaterial { fn create( @@ -13,17 +16,17 @@ impl CreateMaterial for CoatedDiffuseMaterial { normal_map: Option>, named_materials: &HashMap, loc: &FileLoc, - area: &mut Arena, - ) -> Self { + arena: &mut Arena, + ) -> Result { let reflectance = parameters .get_spectrum_texture("reflectance", None, SpectrumType::Albedo) - .unwrap_or(SpectrumConstantTexture::new(Spectrum::Constant( - ConstantSpectrum::new(0.5), + .unwrap_or(Arc::new(SpectrumTexture::Constant( + SpectrumConstantTexture::new(Spectrum::Constant(ConstantSpectrum::new(0.5))), ))); let u_roughness = parameters .get_float_texture_or_null("uroughness") - .unwrap_or_else(|| parameters.get_float_texture("roughness", 0.5)); + .or_else(|| parameters.get_float_texture("roughness", 0.5)); let v_roughness = parameters .get_float_texture_or_null("vroughness") .unwap_or_else(|| parameters.get_float("roughness", 0.5)); @@ -42,24 +45,27 @@ impl CreateMaterial for CoatedDiffuseMaterial { let albedo = parameters .get_spectrum_texture("albedo", None, SpectrumType::Albedo) .unwrap_or_else(|| { - SpectrumConstantTexture::new(Spectrum::Constant(ConstantSpectrum::new(0.))) + let default_spectrum = Spectrum::Constant(ConstantSpectrum::new(0.)); + SpectrumTexture::Constant(SpectrumConstantTexture::new(default_spectrum)) }); - let displacement = parameters.get_float_texture("displacement"); + let displacement = parameters.get_float_texture("displacement", 0.); let remap_roughness = parameters.get_one_bool("remaproughness", true); - arena.alloc(Self::new( - reflectance, - u_roughness, - v_roughness, - thickness, - albedo, - g, + let specific = CoatedDiffuseMaterial::new( + reflectance.upload(arena), + u_roughness.upload(arena), + v_roughness.upload(arena), + thickness.upload(arena), + albedo.upload(arena), + g.upload(arena), eta, - displacement, - normal_map, + displacement.upload(arena), + normal_map.upload(arena), remap_roughness, max_depth, n_samples, - )); + ); + + Ok(Material::CoatedDiffuse(specific)); } } @@ -73,7 +79,7 @@ impl CreateMaterial for CoatedConductorMaterial { ) -> Result { let interface_u_roughness = parameters .get_float_texture_or_null("interface.uroughness") - .unwrap_or_else(|| parameters.get_float_texture("interface.roughness", 0.)); + .r_else(|| parameters.get_float_texture("interface.roughness", 0.)); let interface_v_roughness = parameters .GetFloatTextureOrNull("interface.vroughness") .unwrap_or_else(|| parameters.get_float_texture("interface.vroughness", 0.)); @@ -123,8 +129,8 @@ impl CreateMaterial for CoatedConductorMaterial { let displacement = parameters.get_float_texture_or_null("displacement"); let remap_roughness = parameters.get_one_bool("remaproughness", true); let material = Self::new( - displacement, - normal_map, + displacement.upload(arena), + normal_map.upload(arena), interface_u_roughness, interface_v_roughness, thickness, diff --git a/src/materials/complex.rs b/src/materials/complex.rs index 53b3839..3af3383 100644 --- a/src/materials/complex.rs +++ b/src/materials/complex.rs @@ -1,76 +1,52 @@ use crate::core::material::CreateMaterial; -use crate::utils::{Arena, FileLoc, TextureParameterDictionary}; -use shared::core::bxdf::HairBxDF; +use crate::utils::{Arena, FileLoc, TextureParameterDictionary, Upload}; +use shared::bxdfs::HairBxDF; +use shared::core::material::Material; use shared::core::spectrum::Spectrum; -use shared::core::texture::{GPUFloatTexture, GPUSpectrumTexture}; +use shared::core::texture::GPUFloatTexture; use shared::materials::complex::*; impl CreateMaterial for HairMaterial { fn create( parameters: &TextureParameterDictionary, normal_map: Option>, - named_materials: &HashMap, + _named_materials: &HashMap, loc: &FileLoc, arena: &mut Arena, - ) -> Result { + ) -> Result { let sigma_a = parameters.get_spectrum_texture_or_null("sigma_a", SpectrumType::Unbounded); let reflectance = parameters .get_spectrum_texture_or_null("reflectance", SpectrumType::Albedo) .or_else(|| parameters.get_spectrum_texture_or_null("color", SpectrumType::Albedo)); let eumelanin = parameters.get_float_texture_or_null("eumelanin"); let pheomelanin = parameters.get_float_texture_or_null("pheomelanin"); - let sigma_a = match ( - sigma_a, - reflectance, - eumelanin.is_some() || pheomelanin.is_some(), - ) { - (Some(s), Some(_), _) => { - warn( - loc, - r#"Ignoring "reflectance" parameter since "sigma_a" was provided."#, - ); - Some(s) - } - (Some(s), _, true) => { - warn( - loc, - r#"Ignoring "eumelanin"/"pheomelanin" parameter since "sigma_a" was provided."#, - ); - Some(s) - } - (Some(s), None, false) => Some(s), + let has_melanin = eumelanin.is_some() || pheomelanin.is_some(); - (None, Some(r), true) => { - warn( - loc, - r#"Ignoring "eumelanin"/"pheomelanin" parameter since "reflectance" was provided."#, - ); - Some(r) - } - (None, Some(r), false) => Some(r), - - (None, None, true) => None, // eumelanin/pheomelanin will be used - - (None, None, false) => Some(SpectrumConstantTexture::new(Spectrum::RGBUnbounded( - RGBUnboundedSpectrum::new(HairBxDF::sigma_a_from_concentration(1.3, 0.0)), - ))), + // Default distribution if nothing is spceified + let sigma_a = if sigma_a.is_none() && color.is_none() && !has_melanin { + let default_rgb = HairBxDF::sigma_a_from_concentration(1.3, 0.0); + Some(Arc::new(SpectrumConstantTexture::new( + Spectrum::RGBUnbounded(RGBUnboundedSpectrum::new(default_rgb)), + ))) + } else { + sigma_a }; + let eta = parameters.get_flot_texture("eta", 1.55); let beta_m = parameters.get_float_texture("beta_m", 0.3); let beta_n = parameters.get_float_texture("beta_n", 0.3); let alpha = parameters.get_float_texture("alpha", 2.); let material = HairMaterial::new( - sigma_a, - reflectance, - eumelanin, - pheomelanin, - eta, - beta_m, - beta_n, - alpha, + sigma_a.upload(arena), + reflectance.upload(arena), + eumelanin.upload(arena), + pheomelanin.upload(arena), + eta.upload(arena), + beta_m.upload(arena), + beta_n.upload(arena), + alpha.upload(arena), ); - arena.alloc(material); return material; } } diff --git a/src/materials/conductor.rs b/src/materials/conductor.rs index 9de9429..75911f3 100644 --- a/src/materials/conductor.rs +++ b/src/materials/conductor.rs @@ -1,16 +1,11 @@ -use crate::core::bssrdf::BSSRDF; -use crate::core::bxdf::{ - BSDF, BxDF, CoatedConductorBxDF, CoatedDiffuseBxDF, ConductorBxDF, DielectricBxDF, DiffuseBxDF, - HairBxDF, -}; use crate::core::image::Image; -use crate::core::material::{Material, MaterialEvalContext, MaterialTrait}; 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::core::bsdf::BSDF; +use shared::core::bssrdf::BSSRDF; +use shared::core::material::{MaterialEvalContext, MaterialTrait}; +use shared::core::texture::{GPUFloatTexture, GPUSpectrumTexture, TextureEvaluator}; +use shared::spectra::SampledWavelengths; +use shared::utils::Ptr; #[repr(C)] #[derive(Clone, Copy, Debug)] @@ -22,7 +17,7 @@ pub struct ConductorMaterial { pub u_roughness: Ptr, pub v_roughness: Ptr, pub remap_roughness: bool, - pub normal_map: *const Image, + pub normal_map: Ptr, } impl MaterialTrait for ConductorMaterial { diff --git a/src/materials/dielectric.rs b/src/materials/dielectric.rs index b2a97fa..743db2d 100644 --- a/src/materials/dielectric.rs +++ b/src/materials/dielectric.rs @@ -1,15 +1,12 @@ -use crate::core::bssrdf::BSSRDF; -use crate::core::bxdf::{ - BSDF, BxDF, CoatedConductorBxDF, CoatedDiffuseBxDF, ConductorBxDF, DielectricBxDF, DiffuseBxDF, - HairBxDF, -}; use crate::core::image::Image; -use crate::core::material::{Material, MaterialEvalContext, MaterialTrait}; 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::math::clamp; +use shared::core::bsdf::BSDF; +use shared::core::bssrdf::BSSRDF; +use shared::core::bxdf::BxDF; +use shared::core::material::{MaterialEvalContext, MaterialTrait}; +use shared::core::spectrum::{Spectrum, SpectrumTrait}; +use shared::core::texture::{GPUFloatTexture, TextureEvaluator}; +use shared::spectra::SampledWavelengths; use shared::utils::Ptr; #[repr(C)] diff --git a/src/materials/diffuse.rs b/src/materials/diffuse.rs index e588d30..d26661d 100644 --- a/src/materials/diffuse.rs +++ b/src/materials/diffuse.rs @@ -1,16 +1,12 @@ -use crate::core::bssrdf::BSSRDF; -use crate::core::bxdf::{ - BSDF, BxDF, CoatedConductorBxDF, CoatedDiffuseBxDF, ConductorBxDF, DielectricBxDF, DiffuseBxDF, - HairBxDF, -}; use crate::core::image::Image; -use crate::core::material::{Material, MaterialEvalContext, MaterialTrait}; 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::core::bsdf::BSDF; +use shared::core::bssrdf::BSSRDF; +use shared::core::bxdf::BxDF; +use shared::core::material::{MaterialEvalContext, MaterialTrait}; +use shared::core::texture::{GPUFloatTexture, GPUSpectrumTexture, TextureEvaluator}; +use shared::spectra::SampledWavelengths; +use shared::utils::Ptr; #[repr(C)] #[derive(Clone, Copy, Debug)] diff --git a/src/materials/mix.rs b/src/materials/mix.rs index d17f215..a06112b 100644 --- a/src/materials/mix.rs +++ b/src/materials/mix.rs @@ -1,17 +1,12 @@ -use crate::core::bssrdf::BSSRDF; -use crate::core::bxdf::{ - BSDF, BxDF, CoatedConductorBxDF, CoatedDiffuseBxDF, ConductorBxDF, DielectricBxDF, DiffuseBxDF, - HairBxDF, -}; use crate::core::image::Image; -use crate::core::material::{Material, MaterialEvalContext, MaterialTrait}; 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::hash::hash_float; -use crate::utils::math::clamp; +use shared::core::bsdf::BSDF; +use shared::core::bssrdf::BSSRDF; +use shared::core::material::{Material, MaterialEvalContext, MaterialTrait}; +use shared::core::texture::{GPUFloatTexture, TextureEvaluator}; +use shared::spectra::SampledWavelengths; +use shared::utils::Ptr; +use shared::utils::hash::hash_float; #[repr(C)] #[derive(Clone, Copy, Debug)] diff --git a/src/samplers/halton.rs b/src/samplers/halton.rs index 4180708..c5d94ea 100644 --- a/src/samplers/halton.rs +++ b/src/samplers/halton.rs @@ -1,14 +1,15 @@ -use crate::core::sampler::SamplerFactory; +use crate::core::sampler::CreateSampler; use crate::utils::{FileLoc, ParameterDictionary}; use shared::core::geometry::Point2i; use shared::core::sampler::{HaltonSampler, RandomizeStrategy}; -impl SamplerFactory for HaltonSampler { +impl CreateSampler for HaltonSampler { fn create( params: &ParameterDictionary, full_res: Point2i, loc: &FileLoc, - ) -> Result { + arena: &mut Arena, + ) -> Result { let options = get_options(); let nsamp = options .quick_render @@ -32,6 +33,8 @@ impl SamplerFactory for HaltonSampler { } }; - Ok(HaltonSampler::new(nsamp as u32, full_res, s, seed as u64)) + let sampler = HaltonSampler::new(nsamp as u32, full_res, s, seed as u64); + arena.alloc(sampler); + Ok(sampler) } } diff --git a/src/samplers/independent.rs b/src/samplers/independent.rs index 3b23cc5..6300155 100644 --- a/src/samplers/independent.rs +++ b/src/samplers/independent.rs @@ -1,14 +1,15 @@ -use crate::core::sampler::SamplerFactory; +use crate::core::sampler::CreateSampler; use crate::utils::{FileLoc, ParameterDictionary}; use shared::core::geometry::Point2i; use shared::core::sampler::IndependentSampler; -impl SamplerFactory for IndependentSampler { +impl CreateSampler for IndependentSampler { fn create( params: &ParameterDictionary, _full_res: Point2i, _loc: &FileLoc, - ) -> Result { + arena: &mut Arena, + ) -> Result { let options = get_options(); let nsamp = options .quick_render @@ -16,6 +17,8 @@ impl SamplerFactory for IndependentSampler { .or(options.pixel_samples) .unwrap_or_else(|| params.get_one_int("pixelsamples", 16)); let seed = params.get_one_int("seed", options.seed); - Ok(Self::new(nsamp as usize, seed as u64)) + let sampler = Self::new(nsamp as usize, seed as u64); + arena.alloc(sampler); + Ok(sampler) } } diff --git a/src/samplers/sobol.rs b/src/samplers/sobol.rs index 0ebea81..144a1fd 100644 --- a/src/samplers/sobol.rs +++ b/src/samplers/sobol.rs @@ -1,14 +1,15 @@ -use crate::core::sampler::SamplerFactory; +use crate::core::sampler::CreateSampler; use crate::utils::{FileLoc, ParameterDictionary}; use shared::core::geometry::Point2i; use shared::core::sampler::{PaddedSobolSampler, RandomizeStrategy, SobolSampler, ZSobolSampler}; -impl SamplerFactory for SobolSampler { +impl CreateSampler for SobolSampler { fn create( params: &ParameterDictionary, full_res: Point2i, loc: &FileLoc, - ) -> Result { + arena: &mut Arena, + ) -> Result { let options = get_options(); let nsamp = options .quick_render @@ -26,16 +27,19 @@ impl SamplerFactory for SobolSampler { } }; - Ok(Self::new(nsamp as usize, full_res, s, Some(seed as u64))) + let sampler = Self::new(nsamp as usize, full_res, s, Some(seed as u64)); + arena.alloc(sampler); + Ok(sampler) } } -impl SamplerFactory for PaddedSobolSampler { +impl CreateSampler for PaddedSobolSampler { fn create( params: &ParameterDictionary, _full_res: Point2i, loc: &FileLoc, - ) -> Result { + arena: &mut Arena, + ) -> Result { let options = get_options(); let nsamp = options .quick_render @@ -56,16 +60,19 @@ impl SamplerFactory for PaddedSobolSampler { } }; - Ok(Self::new(nsamp as u32, s, Some(seed as u64))) + let sampler = Self::new(nsamp as u32, s, Some(seed as u64)); + arena.alloc(sampler); + Ok(sampler) } } -impl SamplerFactory for ZSobolSampler { +impl CreateSampler for ZSobolSampler { fn create( params: &ParameterDictionary, full_res: Point2i, loc: &FileLoc, - ) -> Result { + arena: &mut Arena, + ) -> Result { let options = get_options(); let nsamp = options .quick_render @@ -86,11 +93,8 @@ impl SamplerFactory for ZSobolSampler { } }; - Ok(ZSobolSampler::new( - nsamp as u32, - full_res, - s, - Some(seed as u64), - )) + let sampler = ZSobolSampler::new(nsamp as u32, full_res, s, Some(seed as u64)); + arena.alloc(sampler); + Ok(sampler) } } diff --git a/src/samplers/stratified.rs b/src/samplers/stratified.rs index 4f992b3..e84073b 100644 --- a/src/samplers/stratified.rs +++ b/src/samplers/stratified.rs @@ -1,14 +1,16 @@ -use crate::core::sampler::SamplerFactory; +use crate::Arena; +use crate::core::sampler::CreateSampler; use crate::utils::{FileLoc, ParameterDictionary}; use shared::core::geometry::{FileLoc, ParameterDictionary}; use shared::core::sampler::StratifiedSampler; -impl SamplerFactory for StratifiedSampler { +impl CreateSampler for StratifiedSampler { fn create( params: &ParameterDictionary, _full_res: Point2i, _loc: &FileLoc, - ) -> Result { + arena: &mut Arena, + ) -> Result { let options = get_options(); let jitter = params.get_one_bool("jitter", true); let (x_samples, y_samples) = if options.quick_render { @@ -25,11 +27,14 @@ impl SamplerFactory for StratifiedSampler { ) }; let seed = params.get_one_int("seed", options.seed); - Ok(Self::new( + let sampler = Self::new( x_samples as u32, y_samples as u32, Some(seed as u64), jitter, - )) + ); + arena.aloc(sampler); + + Ok(sampler) } } diff --git a/src/shapes/bilinear.rs b/src/shapes/bilinear.rs index 3e561a4..5eeb045 100644 --- a/src/shapes/bilinear.rs +++ b/src/shapes/bilinear.rs @@ -1,8 +1,9 @@ -use crate::core::shape::{ALL_BILINEAR_MESHES, CreateMesh, CreateShape}; +use crate::core::shape::{ALL_BILINEAR_MESHES, CreateMesh}; use crate::core::texture::FloatTexture; -use crate::shapes::mesh::Mesh; use crate::utils::{Arena, FileLoc, ParameterDictionary}; +use log::warn; use shared::shapes::BilinearPatchShape; +use shared::shapes::mesh::Mesh; use shared::utils::Transform; use std::collections::HashMap; diff --git a/src/shapes/curves.rs b/src/shapes/curves.rs index 3aa3191..514a1ee 100644 --- a/src/shapes/curves.rs +++ b/src/shapes/curves.rs @@ -1,13 +1,15 @@ +use crate::Arena; use crate::core::shape::CreateShape; use crate::core::texture::FloatTexture; use crate::utils::{FileLoc, ParameterDictionary}; -use rayon::iter::split; use shared::core::geometry::Normal3f; -use shared::shapes::{CurveCommon, CurveShape, CurveType, CylinderShape}; +use shared::shapes::{CurveCommon, CurveShape, CurveType}; use shared::utils::Transform; use shared::utils::splines::{ cubic_bspline_to_bezier, elevate_quadratic_bezier_to_cubic, quadratic_bspline_to_bezier, }; + +use log::warn; use std::collections::HashMap; pub fn create_curve( @@ -58,6 +60,7 @@ impl CreateShape for CurveShape { parameters: ParameterDictionary, float_textures: HashMap, loc: FileLoc, + arena: &mut Arena, ) -> Result, String> { let width = parameters.get_one_float("width", 1.0); let width0 = parameters.get_one_float("width0", width); diff --git a/src/shapes/mesh.rs b/src/shapes/mesh.rs index 9420c01..ff0dde3 100644 --- a/src/shapes/mesh.rs +++ b/src/shapes/mesh.rs @@ -1,10 +1,9 @@ -use crate::utils::FileLoc; +use anyhow::{Context, Result as AnyResult, bail}; use ply_rs::parser::Parser; use ply_rs::ply::{DefaultElement, Property}; use shared::utils::Transform; use shared::utils::mesh::{BilinearPatchMesh, TriangleMesh}; use shared::utils::sampling::DevicePiecewiseConstant2D; -use std::collections::HashMap; use std::fs::File; use std::path::Path; @@ -18,7 +17,7 @@ pub struct TriQuadMesh { pub quad_indices: Vec, } -fn get_float(elem: &DefaultElement, key: &str) -> Result { +fn get_float(elem: &DefaultElement, key: &str) -> AnyResult { match elem.get(key) { Some(Property::Float(v)) => Ok(*v), Some(Property::Double(v)) => Ok(*v as f32), @@ -27,7 +26,7 @@ fn get_float(elem: &DefaultElement, key: &str) -> Result { } } -fn get_int(elem: &DefaultElement, key: &str) -> Result { +fn get_int(elem: &DefaultElement, key: &str) -> AnyResult { match elem.get(key) { Some(Property::Int(v)) => Ok(*v), Some(Property::UInt(v)) => Ok(*v as i32), @@ -48,7 +47,7 @@ fn get_float_any(elem: &DefaultElement, keys: &[&str]) -> Option { None } -fn get_list_uint(elem: &DefaultElement, key: &str) -> Result> { +fn get_list_uint(elem: &DefaultElement, key: &str) -> AnyResult> { match elem.get(key) { Some(Property::List_Int(vec)) => Ok(vec.iter().map(|&x| x as u32).collect()), Some(Property::List_UInt(vec)) => Ok(vec.clone()), @@ -59,10 +58,14 @@ fn get_list_uint(elem: &DefaultElement, key: &str) -> Result> { } impl TriQuadMesh { - pub fn read_ply>(filename: P) -> Result { - let filename_display = filename.as_ref().display().to_string(); + pub fn read_ply>(filename: P) -> AnyResult { + let path = filename.as_ref(); + let filename_display = path.display().to_string(); + let mut f = File::open(&filename) .with_context(|| format!("Couldn't open PLY file \"{}\"", filename_display))?; + + // Going to ply-rs let p = Parser::::new(); let ply = p .read_ply(&mut f) @@ -78,13 +81,13 @@ impl TriQuadMesh { let has_normal = first.contains_key("nx") && first.contains_key("ny") && first.contains_key("nz"); for v_elem in vertices { - // Read Position (Required) + // Read Position let x = get_float(v_elem, "x")?; let y = get_float(v_elem, "y")?; let z = get_float(v_elem, "z")?; mesh.p.push([x, y, z]); - // Read Normal (Optional) + // Read Normal if has_normal { let nx = get_float(v_elem, "nx").unwrap_or(0.0); let ny = get_float(v_elem, "ny").unwrap_or(0.0); @@ -343,6 +346,12 @@ pub struct BilinearPatchMeshHost { _storage: Box, } +#[derive(Debug, Clone, Copy)] +pub struct TriangleMeshHost { + pub view: BilinearPatchMesh, + _storage: Box, +} + impl Deref for TriangleMeshHost { type Target = TriangleMesh; fn deref(&self) -> &Self::Target { diff --git a/src/shapes/mod.rs b/src/shapes/mod.rs index 1f14aac..483f933 100644 --- a/src/shapes/mod.rs +++ b/src/shapes/mod.rs @@ -7,7 +7,7 @@ pub mod mesh; pub mod sphere; pub mod triangle; -pub use bilinear::*; +// pub use bilinear::*; pub use mesh::*; use std::sync::{Arc, Mutex}; diff --git a/src/shapes/triangle.rs b/src/shapes/triangle.rs index 2bc32b7..bf29f96 100644 --- a/src/shapes/triangle.rs +++ b/src/shapes/triangle.rs @@ -2,6 +2,7 @@ use crate::core::shape::{ALL_TRIANGLE_MESHES, CreateShape}; use crate::core::texture::FloatTexture; use crate::shapes::mesh::TriangleMeshHost; use crate::utils::{Arena, FileLoc, ParameterDictionary}; +use log::warn; use shared::shapes::TriangleShape; use shared::utils::Transform; @@ -9,8 +10,10 @@ impl CreateShape for TriangleShape { fn create( name: &str, render_from_object: Transform, + _object_from_render: Transform, reverse_orientation: bool, parameters: ParameterDictionary, + _float_texture: HashMap, loc: FileLoc, arena: &mut Arena, ) -> Result { diff --git a/src/spectra/colorspace.rs b/src/spectra/colorspace.rs index d47212f..77c9810 100644 --- a/src/spectra/colorspace.rs +++ b/src/spectra/colorspace.rs @@ -1,49 +1,52 @@ use super::DenselySampledSpectrumBuffer; +use shared::core::color::RGBToSpectrumTable; use shared::core::geometry::Point2f; -use shared::spectra::{RGBColorSpace, RGBToSpectrumTable}; -use shared::utils::SquareMatrix; +use shared::spectra::RGBColorSpace; +use shared::utils::math::SquareMatrix; pub struct RGBColorSpaceData { - _illuminant: DenselySampledBuffer, + _illuminant: DenselySampledSpectrumBuffer, pub view: RGBColorSpace, } -pub fn new( - r: Point2f, - g: Point2f, - b: Point2f, - illuminant: DenselySampledBuffer, - rgb_to_spectrum_table: *const RGBToSpectrumTable, -) -> Self { - let w_xyz: XYZ = illuminant.to_xyz(); - let w = w_xyz.xy(); - let r_xyz = XYZ::from_xyy(r, Some(1.0)); - let g_xyz = XYZ::from_xyy(g, Some(1.0)); - let b_xyz = XYZ::from_xyy(b, Some(1.0)); - let rgb_values = [ - [r_xyz.x(), g_xyz.x(), b_xyz.x()], - [r_xyz.y(), g_xyz.y(), b_xyz.y()], - [r_xyz.z(), g_xyz.z(), b_xyz.z()], - ]; - let rgb = SquareMatrix::new(rgb_values); - let c: RGB = rgb.inverse()? * w_xyz; - let xyz_from_rgb = rgb * SquareMatrix::diag(&[c.r, c.g, c.b]); - let rgb_from_xyz = xyz_from_rgb - .inverse() - .expect("XYZ from RGB matrix is singular"); - let view = RGBColorSpace { - r, - g, - b, - w, - illuminant: illuminant_buffer.view, // Extract view - xyz_from_rgb, - rgb_from_xyz, - rgb_to_spectrum_table: table_ptr, - }; - Self { - _illuminant: illuminant_buffer, - view, +impl RGBColorSpaceData { + pub fn new( + r: Point2f, + g: Point2f, + b: Point2f, + illuminant: DenselySampledSpectrumBuffer, + rgb_to_spectrum_table: *const RGBToSpectrumTable, + ) -> Self { + let w_xyz: XYZ = illuminant.to_xyz(); + let w = w_xyz.xy(); + let r_xyz = XYZ::from_xyy(r, Some(1.0)); + let g_xyz = XYZ::from_xyy(g, Some(1.0)); + let b_xyz = XYZ::from_xyy(b, Some(1.0)); + let rgb_values = [ + [r_xyz.x(), g_xyz.x(), b_xyz.x()], + [r_xyz.y(), g_xyz.y(), b_xyz.y()], + [r_xyz.z(), g_xyz.z(), b_xyz.z()], + ]; + let rgb = SquareMatrix::new(rgb_values); + let c: RGB = rgb.inverse()? * w_xyz; + let xyz_from_rgb = rgb * SquareMatrix::diag(&[c.r, c.g, c.b]); + let rgb_from_xyz = xyz_from_rgb + .inverse() + .expect("XYZ from RGB matrix is singular"); + let view = RGBColorSpace { + r, + g, + b, + w, + illuminant: illuminant_buffer.view, // Extract view + xyz_from_rgb, + rgb_from_xyz, + rgb_to_spectrum_table: table_ptr, + }; + Self { + _illuminant: illuminant_buffer, + view, + } } } diff --git a/src/spectra/data.rs b/src/spectra/data.rs index ebfbd56..48f2949 100644 --- a/src/spectra/data.rs +++ b/src/spectra/data.rs @@ -1,16 +1,9 @@ -use crate::spectra::{DenselySampledSpectrumBuffer, SpectrumData}; +use crate::spectra::DenselySampledSpectrumBuffer; use shared::Float; -use shared::core::cie::{CIE_S_LAMBDA, CIE_S0, CIE_S1, CIE_S2, CIE_X, CIE_Y, CIE_Z, N_CIES}; -use shared::spectra::cie::{CIE_S_LAMBDA, CIE_S0, CIE_S1, CIE_S2, N_CIES}; -use shared::spectra::{ - BlackbodySpectrum, DenselySampledSpectrum, LAMBDA_MAX, LAMBDA_MIN, PiecewiseLinearSpectrum, - Spectrum, -}; -use shared::utils::math::square; +use shared::core::spectrum::Spectrum; +use shared::spectra::PiecewiseLinearSpectrum; -use once_cell::sync::Lazy; - -fn create_cie_buffer(data: &[Float]) -> Spectrum { +pub fn create_cie_buffer(data: &[Float]) -> Spectrum { let buffer = PiecewiseLinearSpectrum::from_interleaved(data, false); let spec = Spectrum::Piecewise(buffer.view); DenselySampledSpectrumBuffer::from_spectrum(spec) diff --git a/src/spectra/dense.rs b/src/spectra/dense.rs index 6359d82..06c33e6 100644 --- a/src/spectra/dense.rs +++ b/src/spectra/dense.rs @@ -1,15 +1,15 @@ -use crate::core::pbrt::Float; +use shared::Float; use shared::spectra::{DenselySampledSpectrum, SampledSpectrum}; pub struct DenselySampledSpectrumBuffer { - pub view: DenselySampledSpectrum, + pub device: DenselySampledSpectrum, _storage: Vec, } impl std::ops::Deref for DenselySampledSpectrumBuffer { type Target = DenselySampledSpectrum; fn deref(&self) -> &Self::Target { - &self.view + &self.device } } @@ -21,7 +21,7 @@ impl DenselySampledSpectrumBuffer { values: values.as_ptr(), }; Self { - view, + device: view, _storage: values, } } diff --git a/src/spectra/mod.rs b/src/spectra/mod.rs index d6414fe..a9a1287 100644 --- a/src/spectra/mod.rs +++ b/src/spectra/mod.rs @@ -1,20 +1,27 @@ -use shared::cie::{CIE_D65, CIE_X, CIE_Y, CIE_Z}; +use crate::spectra::colorspace::RGBColorSpaceData; +use once_cell::unsync::Lazy; +use shared::core::geometry::Point2f; use shared::core::spectrum::Spectrum; -use shared::spectra::RGBToSpectrumTable; -use std::sync::Lazy; +use shared::spectra::cie::{CIE_D65, CIE_X, CIE_Y, CIE_Z}; +use shared::spectra::{RGBColorSpace, RGBToSpectrumTable, StandardColorSpaces}; +use std::sync::Arc; pub mod colorspace; pub mod data; pub mod dense; pub mod piecewise; -use data; +// pub use data; pub use dense::DenselySampledSpectrumBuffer; -static CIE_X_DATA: Lazy = Lazy::new(|| data::create_cie_buffer(&CIE_X)); -static CIE_Y_DATA: Lazy = Lazy::new(|| data::create_cie_buffer(&CIE_Y)); -static CIE_Z_DATA: Lazy = Lazy::new(|| data::create_cie_buffer(&CIE_Z)); -static CIE_D65_DATA: Lazy = Lazy::new(|| data::create_cie_buffer(&CIE_D65)); +static CIE_X_DATA: Lazy = + Lazy::new(|| data::create_cie_buffer(&CIE_X)); +static CIE_Y_DATA: Lazy = + Lazy::new(|| data::create_cie_buffer(&CIE_Y)); +static CIE_Z_DATA: Lazy = + Lazy::new(|| data::create_cie_buffer(&CIE_Z)); +static CIE_D65_DATA: Lazy = + Lazy::new(|| data::create_cie_buffer(&CIE_D65)); pub fn cie_x() -> Spectrum { Spectrum::DenselySampled(CIE_X_DATA.view) @@ -43,7 +50,7 @@ pub fn get_spectra_context() -> StandardSpectra { } pub static SRGB_DATA: Lazy = Lazy::new(|| { - let illum = DenselySampledBuffer::new( + let illum = DenselySampledSpectrumBuffer::new( D65_BUFFER.view.lambda_min, D65_BUFFER.view.lambda_max, D65_BUFFER._storage.clone(), @@ -59,7 +66,7 @@ pub static SRGB_DATA: Lazy = Lazy::new(|| { }); pub static DCI_P3: Lazy = Lazy::new(|| { - let illum = DenselySampledBuffer::new( + let illum = DenselySampledSpectrumBuffer::new( D65_BUFFER.view.lambda_min, D65_BUFFER.view.lambda_max, D65_BUFFER._storage.clone(), @@ -73,7 +80,7 @@ pub static DCI_P3: Lazy = Lazy::new(|| { }); pub static REC2020: Lazy> = Lazy::new(|| { - let illum = DenselySampledBuffer::new( + let illum = DenselySampledSpectrumBuffer::new( D65_BUFFER.view.lambda_min, D65_BUFFER.view.lambda_max, D65_BUFFER._storage.clone(), diff --git a/src/textures/image.rs b/src/textures/image.rs index bda13ab..073b98b 100644 --- a/src/textures/image.rs +++ b/src/textures/image.rs @@ -1,5 +1,18 @@ -use crate::core::texture::get_texture_cache; -use crate::utils::{FileLoc, TextureParameterDictionary}; +use crate::core::texture::{TexInfo, get_texture_cache}; +use crate::utils::mipmap::{MIPMap, MIPMapFilterOptions}; +use shared::Float; +use shared::core::color::ColorEncoding; +use shared::core::color::RGB; +use shared::core::geometry::Vector2f; +use shared::core::image::{WrapMode, WrapMode2D}; +use shared::core::texture::{SpectrumType, TextureEvalContext, TextureMapping2D}; +use shared::spectra::{ + RGBAlbedoSpectrum, RGBIlluminantSpectrum, RGBUnboundedSpectrum, SampledSpectrum, + SampledWavelengths, +}; +use std::path::Path; +use std::sync::Arc; +// use crate::utils::{FileLoc, TextureParameterDictionary}; #[derive(Clone, Debug)] pub struct ImageTextureBase { diff --git a/src/textures/mix.rs b/src/textures/mix.rs index 1f72fc9..187eeb3 100644 --- a/src/textures/mix.rs +++ b/src/textures/mix.rs @@ -1,8 +1,8 @@ -use crate::core::texture::{ - FloatTexture, FloatTextureTrait, SpectrumTexture, SpectrumTextureTrait, -}; +use crate::core::texture::FloatTexture; use crate::utils::{Arena, FileLoc, TextureParameterDictionary}; +use shared::Float; use shared::core::geometry::Vector3f; +use shared::core::texture::{FloatTextureTrait, SpectrumTextureTrait, TextureEvalContext}; use shared::utils::Transform; use std::sync::Arc; diff --git a/src/textures/ptex.rs b/src/textures/ptex.rs index 115b7cc..e6b2cb2 100644 --- a/src/textures/ptex.rs +++ b/src/textures/ptex.rs @@ -1,21 +1,31 @@ -use crate::utils::{Arena, FileLoc, TextureParameterDictionary}; +// use crate::utils::{Arena, FileLoc, TextureParameterDictionary}; +use shared::core::color::{ColorEncoding, RGB}; +use shared::core::spectrum::SpectrumTextureTrait; use shared::core::texture::{SpectrumType, TextureEvalContext}; -use shared::spectra::ColorEncoding; -use shared::spectra::color::RGB; +use shared::Float; +use shared::spectra::{SampledSpectrum, SampledWavelengths}; +use image_rs::imageops::FilterType; use ptex::Cache; use ptex_sys::ffi; use std::sync::OnceLock; -static PTEX_CACHE: OnceLock = OnceLock::new(); +struct PtexCache(Cache); +unsafe impl Sync for PtexCache {} +unsafe impl Send for PtexCache {} + +static PTEX_CACHE: OnceLock = OnceLock::new(); fn get_ptex_cache() -> &'static Cache { - PTEX_CACHE.get_or_init(|| { - let max_files = 100; - let max_mem = 1 << 32; - let premultiply = true; - Cache::new(max_files, max_mem, premultiply) - }) + PTEX_CACHE + .get_or_init(|| { + let max_files = 100; + let max_mem = 1 << 32; + let premultiply = true; + let cache = Cache::new(max_files, max_mem, premultiply); + PtexCache(cache) + }) + .0 } #[derive(Debug, Clone)] @@ -57,10 +67,11 @@ impl PtexTextureBase { } } - pub fn sample_texture(ctx: &TextureEvalContext) -> Option { + pub fn sample_texture(&self, ctx: &TextureEvalContext) -> Option { if !self.valid { return None; } + let mut result = [0.0; 3]; let nc = self.ptex_handle.eval( &mut result, @@ -82,18 +93,16 @@ impl PtexTextureBase { }; let nc = texture.num_channels(); - let mut result = [0.0f32; 4]; + let mut result = [0.0; 4]; unsafe { let opts = ffi::PtexFilter_Options { - filter: ffi::FilterType::f_bspline, + filter: FilterType::f_bspline, lerp: 1, sharpness: 0.0, noedgeblend: 0, __structSize: std::mem::size_of::() as i32, }; - // Get the raw filter from the low-level FFI - // Assuming your 'texture' can provide the raw C++ pointer let filter_ptr = ffi::PtexFilter_getFilter(texture.as_ptr(), &opts); if filter_ptr.is_null() { return None; @@ -102,7 +111,7 @@ impl PtexTextureBase { // Evaluate (*filter_ptr).eval( result.as_mut_ptr(), - 0, // first channel + 0, texture.num_channels(), ctx.face_index as i32, ctx.uv[0], @@ -113,7 +122,7 @@ impl PtexTextureBase { ctx.dvdy, ); - // Crucial: Manually release the C++ object + // Manually release the C++ object (*filter_ptr).release(); } diff --git a/src/textures/scaled.rs b/src/textures/scaled.rs index 247fca7..e9991b0 100644 --- a/src/textures/scaled.rs +++ b/src/textures/scaled.rs @@ -1,10 +1,10 @@ -use crate::core::texture::{ - FloatTexture, FloatTextureTrait, SpectrumTexture, SpectrumTextureTrait, -}; +use crate::core::texture::{FloatTexture, SpectrumTexture}; use crate::utils::{Arena, FileLoc, TextureParameterDictionary}; -use shared::core::texture::TextureEvalContext; +use shared::core::texture::{FloatTextureTrait, SpectrumTextureTrait, TextureEvalContext}; use shared::spectra::{SampledSpectrum, SampledWavelengths}; use shared::utils::Transform; +use shared::Float; +use std::sync::Arc; #[derive(Debug, Clone)] pub struct FloatScaledTexture { diff --git a/src/utils/arena.rs b/src/utils/arena.rs index a8b01d7..225a078 100644 --- a/src/utils/arena.rs +++ b/src/utils/arena.rs @@ -1,16 +1,16 @@ use crate::core::image::Image; -use crate::core::texture::FloatTexture; +use crate::core::texture::{FloatTexture, SpectrumTexture}; use crate::utils::sampling::PiecewiseConstant2D; -use shared::Float; use shared::core::color::RGBToSpectrumTable; -use shared::core::image::DeviceImage; use shared::core::shape::Shape; -use shared::core::texture::GPUFloatTexture; +use shared::core::spectrum::Spectrum; +use shared::core::texture::{GPUFloatTexture, GPUSpectrumTexture}; use shared::spectra::{RGBColorSpace, StandardColorSpaces}; use shared::textures::*; use shared::utils::Ptr; -use shared::utils::sampling::DevicePiecewiseConstant2D; +use shared::utils::sampling::{DevicePiecewiseConstant1D, DevicePiecewiseConstant2D}; use std::alloc::Layout; +use std::sync::Arc; pub struct Arena { buffer: Vec<(*mut u8, Layout)>, @@ -26,7 +26,6 @@ impl Arena { let ptr = unsafe { self.alloc_unified(layout) } as *mut T; - // Write the value unsafe { ptr.write(value); } @@ -63,7 +62,7 @@ impl Arena { 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); + let result = unsafe { cudaMallocManaged(&mut ptr, size, cudaMemAttachGlobal) }; if result != cudaError::cudaSuccess { panic!("cudaMallocManaged failed: {:?}", result); @@ -75,7 +74,6 @@ impl Arena { #[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 @@ -86,7 +84,7 @@ impl Arena { } } -impl Drop for UnifiedArena { +impl Drop for Arena { fn drop(&mut self) { for (ptr, layout) in self.allocations.drain(..) { unsafe { @@ -119,12 +117,69 @@ impl Upload for Shape { impl Upload for Image { type Target = Image; fn upload(&self, arena: &mut Arena) -> Ptr { - let pixels_ptr = arena.alloc_slice(&self.storage_as_slice()); - let device_img = Image { - base: self.base, - pixels: pixels_ptr, + arena.alloc(self.clone()) + } +} + +impl Upload for Spectrum { + type Target = Spectrum; + fn upload(&self, arena: &mut Arena) -> Ptr { + arena.alloc(self.clone()) + } +} + +impl Upload for SpectrumTexture { + type Target = GPUSpectrumTexture; + fn upload(&self, arena: &mut Arena) -> Ptr { + let gpu_variant = match self { + SpectrumTexture::Constant(tex) => GPUSpectrumTexture::Constant(tex.clone()), + SpectrumTexture::Checkerboard(tex) => GPUSpectrumTexture::Checkerboard(tex.clone()), + SpectrumTexture::Dots(tex) => GPUSpectrumTexture::Dots(tex.clone()), + SpectrumTexture::Image(tex) => GPUSpectrumTexture::Image(tex.clone()), + SpectrumTexture::Bilerp(tex) => GPUSpectrumTexture::Bilerp(tex.clone()), + SpectrumTexture::Scaled(tex) => { + let child_ptr = tex.texture.upload(arena); + + let gpu_scaled = GPUFloatScaledTexture { + tex: child_ptr, + scale: tex.scale.upload(arena), + }; + GPUSpectrumTexture::Scaled(gpu_scaled) + } + SpectrumTexture::Ptex(tex) => GPUSpectrumTexture::Ptex(tex.clone()), + SpectrumTexture::Marble(tex) => GPUSpectrumTexture::Marble(tex.clone()), + SpectrumTexture::RGBConstant(tex) => GPUSpectrumTexture::RGBConstant(tex.clone()), + SpectrumTexture::RGBReflectanceConstant(tex) => { + GPUSpectrumTexture::RGBReflectanceConstant(tex.clone()) + } + SpectrumTexture::Mix(tex) => { + let tex1_ptr = tex.tex1.upload(arena); + let tex2_ptr = tex.tex2.upload(arena); + let amount_ptr = tex.amount.upload(arena); + + let gpu_mix = GPUSpectrumMixTexture { + tex1: tex1_ptr, + tex2: tex2_ptr, + amount: amount_ptr, + }; + + GPUSpectrumTexture::Mix(gpu_mix) + } + SpectrumTexture::DirectionMix(tex) => { + let tex1_ptr = tex.tex1.upload(arena); + let tex2_ptr = tex.tex2.upload(arena); + + let gpu_mix = GPUSpectrumDirectionMixTexture { + tex1: tex1_ptr, + tex2: tex2_ptr, + dir: tex.dir, + }; + + GPUSpectrumTexture::DirectionMix(gpu_mix) + } }; - arena.alloc(device_img) + + arena.alloc(gpu_variant) } } @@ -287,3 +342,22 @@ impl Upload for Option { } } } + +impl Upload for Arc { + type Target = T::Target; + + fn upload(&self, arena: &mut Arena) -> Ptr { + (**self).upload(arena) + } +} + +impl Upload for Option> { + type Target = T::Target; + + fn upload(&self, arena: &mut Arena) -> Ptr { + match self { + Some(v) => v.upload(arena), + None => Ptr::null(), + } + } +} diff --git a/src/utils/containers.rs b/src/utils/containers.rs index 3ee3a0f..4966d90 100644 --- a/src/utils/containers.rs +++ b/src/utils/containers.rs @@ -1,6 +1,9 @@ -use crate::core::geometry::{Bounds2i, Point2i}; +use parking_lot::Mutex; +use shared::core::geometry::{Bounds2i, Point2i}; use shared::utils::containers::Array2D; +use std::collections::HashSet; use std::ops::{Deref, DerefMut}; +use std::sync::Arc; pub struct InternCache { cache: Mutex>>, @@ -8,7 +11,7 @@ pub struct InternCache { impl InternCache where - T: Eq + Hash + Clone, + T: Eq + Clone, { pub fn new() -> Self { Self { diff --git a/src/utils/error.rs b/src/utils/error.rs index 57327e0..6c835c5 100644 --- a/src/utils/error.rs +++ b/src/utils/error.rs @@ -1,6 +1,8 @@ -use image_rs::{ImageError as IError, error}; +use image_rs::ImageError as IError; use shared::core::image::PixelFormat; +use std::fmt; use std::sync::Arc; +use thiserror::Error; #[derive(Clone, Debug)] pub struct FileLoc { diff --git a/src/utils/file.rs b/src/utils/file.rs index 7e3e91e..70a91d5 100644 --- a/src/utils/file.rs +++ b/src/utils/file.rs @@ -1,6 +1,6 @@ use shared::Float; use std::fs; -use std::io::{self, BufReader, Error, ErrorKind, Read}; +use std::io::{self, Error, ErrorKind}; use std::path::{Path, PathBuf}; use std::sync::OnceLock; diff --git a/src/utils/math.rs b/src/utils/math.rs index 449b2eb..caf9172 100644 --- a/src/utils/math.rs +++ b/src/utils/math.rs @@ -1,7 +1,7 @@ -use half::f16; use shared::Float; -use shared::utils::DevicePtr; -use shared::utils::math::{DigitPermutation, PRIMES}; +use shared::utils::Ptr; +use shared::utils::hash::hash_buffer; +use shared::utils::math::{DigitPermutation, PRIMES, permutation_element}; pub fn new_digit_permutation(base: u32, seed: u64) -> Vec { let mut n_digits: u32 = 0; @@ -56,7 +56,7 @@ pub fn compute_radical_inverse_permutations(seed: u64) -> (Vec, Vec, message: &str) -> ! { + if let Some(l) = loc { + log::error!("Error: {}: {}", l, message); + } else { + log::error!("Error: {}", message); + } + + std::process::exit(1); +} + #[derive(Debug)] pub struct ParsedParameter { pub type_name: String, @@ -159,10 +166,10 @@ impl PBRTParameter for Point2f { const TYPE_NAME: &'static str = "point2"; const N_PER_ITEM: usize = 2; fn convert(v: &[Self::Raw]) -> Self { - Poin23f::new(v[0], v[1]) + Point2f::new(v[0], v[1]) } fn get_values(param: &ParsedParameter) -> &[Self::Raw] { - params.floats + param.floats } } @@ -242,12 +249,12 @@ impl ParameterDictionary { } pub fn from_array( - mut params0: ParsedParameterVector, - params1: &[ParsedParameter], + mut p0: ParsedParameterVector, + params: &[ParsedParameter], color_space: Option>, ) -> Self { - let n_owned = params.len(); - p0.extend(params1.iter().rev().cloned()); + let n_owned_params = params.len(); + p0.extend(params.iter().rev().cloned()); let dict = Self { params: p0, @@ -327,10 +334,11 @@ impl ParameterDictionary { } } - fn lookup_single(&self, name: &str, def: T) -> T + fn lookup_single(&self, name: &str, default_val: T) -> T where T: PBRTParameter, { + let param = self.params; if param.name == name && param.type_name == T::TYPE_NAME { let values = T::get_values(param); @@ -573,7 +581,7 @@ impl ParameterDictionary { .floats .chunks_exact(3) .map(|chunk| { - let rgb = crate::spectra::RGB::new(chunk[0], chunk[1], chunk[2]); + let rgb = RGB::new(chunk[0], chunk[1], chunk[2]); if rgb.r < 0.0 || rgb.g < 0.0 || rgb.b < 0.0 { panic!( @@ -649,7 +657,7 @@ impl ParameterDictionary { .strings .iter() .map(|s| { - crate::spectra::get_named_spectrum(s) + get_named_spectrum(s) .ok_or(()) .or_else(|_| read_spectrum_from_file(s).map_err(|_| ())) .unwrap_or_else(|_| panic!("{}: {}: unable to read spectrum", param.loc, s)) @@ -661,7 +669,7 @@ impl ParameterDictionary { fn read_spectrum_from_file(filename: &str) -> Result { let fn_key = filename.to_string(); { - let cache = CACHED_SPECTRA.lock().unwrap(); + let cache = SPECTRUM_CACHE.lock().unwrap(); if let Some(s) = cache.get(&fn_key) { return Ok(s.clone()); } @@ -673,7 +681,7 @@ fn read_spectrum_from_file(filename: &str) -> Result { let spectrum = Spectrum::PiecewiseLinear(pls); { - let mut cache = CACHED_SPECTRA.lock().unwrap(); + let mut cache = SPECTRUM_CACHE.lock().unwrap(); cache.insert(fn_key, spectrum.clone()); } diff --git a/src/utils/parser.rs b/src/utils/parser.rs index 3549621..9a115e9 100644 --- a/src/utils/parser.rs +++ b/src/utils/parser.rs @@ -3,9 +3,7 @@ use memmap2::Mmap; use std::collections::HashMap; use std::fs::File; use std::io::{self, Read}; -use std::iter::Peekable; use std::path::{Path, PathBuf}; -use std::str::CharIndices; use std::sync::Arc; use crate::utils::error::FileLoc; @@ -1066,7 +1064,6 @@ impl<'a> SceneParser<'a> { } } - // Handles Keyword "type" "Param list" fn parse_basic_entry(&mut self, mut func: F) -> Result<(), ParserError> where F: FnMut(&mut dyn ParserTarget, &str, &ParsedParameterVector, FileLoc), diff --git a/src/utils/sampling.rs b/src/utils/sampling.rs index 6bdf713..a190427 100644 --- a/src/utils/sampling.rs +++ b/src/utils/sampling.rs @@ -1,8 +1,10 @@ +use crate::core::image::Image; +use crate::utils::Arena; use shared::Float; -use shared::core::geometry::{Bounds2f, Point2f}; -use shared::utils::DevicePtr; +use shared::core::geometry::{Point2i, Vector2f, Vector2i}; +use shared::utils::Ptr; use shared::utils::sampling::{ - AliasTable, DevicePiecewiseConstant1D, DevicePiecewiseConstant2D, PiecewiseLinear2D, + AliasTable, Bin, DevicePiecewiseConstant1D, DevicePiecewiseConstant2D, PiecewiseLinear2D, }; use std::sync::Arc; @@ -27,7 +29,7 @@ impl PiecewiseConstant1D { func: func_ptr, cdf: cdf_ptr, func_integral: self.func_integral, - n: func.len(), + n: self.func.len(), min: self.min, max: self.max, } @@ -184,25 +186,25 @@ impl PiecewiseConstant2D { } } -struct PiecewiseLinear2DStorage { +struct PiecewiseLinear2DStorage { data: Vec, marginal_cdf: Vec, conditional_cdf: Vec, - param_values: [Vec; D], + param_values: [Vec; N], } -pub struct PiecewiseLinear2DHost { - pub view: PiecewiseLinear2D, - _storage: Arc>, +pub struct PiecewiseLinear2DHost { + pub view: PiecewiseLinear2D, + _storage: Arc>, } -impl PiecewiseLinear2DHost { +impl PiecewiseLinear2DHost { pub fn new( data: &[Float], x_size: i32, y_size: i32, - param_res: [usize; D], - param_values: [&[Float]; D], + param_res: [usize; N], + param_values: [&[Float]; N], normalize: bool, build_cdf: bool, ) -> Self { @@ -308,17 +310,17 @@ impl PiecewiseLinear2DHost { }; let storage = Arc::new(PiecewiseLinear2DStorage { - data: host_data, - marginal_cdf: host_marginal, - conditional_cdf: host_conditional, - param_values: host_param_values, + data: new_data, + marginal_cdf, + conditional_cdf, + param_values, }); - let mut final_view = view_struct; + let mut final_view = view; final_view.data = storage.data.as_ptr(); final_view.marginal_cdf = storage.marginal_cdf.as_ptr(); final_view.conditional_cdf = storage.conditional_cdf.as_ptr(); - for i in 0..D { + for i in 0..N { final_view.param_values[i] = storage.param_values[i].as_ptr(); } @@ -341,7 +343,7 @@ impl AliasTableHost { if n == 0 { return Self { view: AliasTable { - bins: ptr::null(), + bins: Ptr::null(), size: 0, }, _storage: Vec::new(),