diff --git a/src/pbrt/gpu/media.cpp b/src/pbrt/gpu/media.cpp index fa46dea822c92a79a8dbfd8ad1ccb8df9afb80e3..64ba3624d88efb2705c613ae18b122ca43fa9be9 100644 --- a/src/pbrt/gpu/media.cpp +++ b/src/pbrt/gpu/media.cpp @@ -168,8 +168,8 @@ void GPUPathIntegrator::SampleMediumInteraction(int depth) { PBRT_DBG("Adding ray to escapedRayQueue pixel index %d depth %d\n", w.pixelIndex, depth); escapedRayQueue->Push(EscapedRayWorkItem{ - T_hat, uniPathPDF, lightPathPDF, lambda, ray.o, ray.d, - w.prevIntrCtx, (int)w.isSpecularBounce, w.pixelIndex}); + ray.o, ray.d, lambda, w.pixelIndex, (int)w.isSpecularBounce, + T_hat, uniPathPDF, lightPathPDF, w.prevIntrCtx}); } } @@ -191,8 +191,9 @@ void GPUPathIntegrator::SampleMediumInteraction(int depth) { "depth %d\n", w.pixelIndex, depth); hitAreaLightQueue->Push(HitAreaLightWorkItem{ - w.areaLight, lambda, T_hat, uniPathPDF, lightPathPDF, Point3f(w.pi), - w.n, w.uv, -ray.d, w.prevIntrCtx, w.isSpecularBounce, w.pixelIndex}); + w.areaLight, Point3f(w.pi), w.n, w.uv, -ray.d, lambda, T_hat, + uniPathPDF, lightPathPDF, w.prevIntrCtx, w.isSpecularBounce, + w.pixelIndex}); } FloatTextureHandle displacement = material.GetDisplacement(); @@ -210,8 +211,8 @@ void GPUPathIntegrator::SampleMediumInteraction(int depth) { using Material = typename std::remove_reference_t; q->Push>(MaterialEvalWorkItem{ ptr, w.pi, w.n, w.ns, w.dpdus, w.dpdvs, w.dndus, w.dndvs, w.uv, - lambda, w.anyNonSpecularBounces, T_hat, uniPathPDF, -ray.d, ray.time, - w.etaScale, w.mediumInterface, w.pixelIndex}); + lambda, w.anyNonSpecularBounces, -ray.d, w.pixelIndex, T_hat, + uniPathPDF, w.mediumInterface, w.etaScale, ray.time}); }; material.Dispatch(enqueue); }); diff --git a/src/pbrt/gpu/optix.cu b/src/pbrt/gpu/optix.cu index d8243a36d4a25ecc7107a7b73e9cda1f8a0ddb16..13d36200f0548a628308971d1c8141865d232bbe 100644 --- a/src/pbrt/gpu/optix.cu +++ b/src/pbrt/gpu/optix.cu @@ -124,8 +124,8 @@ extern "C" __global__ void __raygen__findClosest() { PBRT_DBG("Adding ray to escapedRayQueue ray index %d pixel index %d\n", rayIndex, r.pixelIndex); params.escapedRayQueue->Push(EscapedRayWorkItem{ - r.T_hat, r.uniPathPDF, r.lightPathPDF, r.lambda, ray.o, ray.d, r.prevIntrCtx, - (int)r.isSpecularBounce, r.pixelIndex}); + ray.o, ray.d, r.lambda, r.pixelIndex, (int)r.isSpecularBounce, + r.T_hat, r.uniPathPDF, r.lightPathPDF, r.prevIntrCtx}); } } } @@ -198,7 +198,7 @@ static __forceinline__ __device__ void ProcessClosestIntersection( PBRT_DBG("Enqueuing into medium transition queue: ray index %d pixel index %d \n", rayIndex, r.pixelIndex); Ray newRay = intr.SpawnRay(r.ray.d); - params.nextRayQueue->PushIndirect( + params.nextRayQueue->PushIndirectRay( newRay, r.prevIntrCtx, r.T_hat, r.uniPathPDF, r.lightPathPDF, r.lambda, r.etaScale, r.isSpecularBounce, r.anyNonSpecularBounces, r.pixelIndex); return; @@ -210,8 +210,9 @@ static __forceinline__ __device__ void ProcessClosestIntersection( Ray ray = r.ray; // TODO: intr.wo == -ray.d? params.hitAreaLightQueue->Push(HitAreaLightWorkItem{ - intr.areaLight, r.lambda, r.T_hat, r.uniPathPDF, r.lightPathPDF, intr.p(), intr.n, - intr.uv, intr.wo, r.prevIntrCtx, (int)r.isSpecularBounce, r.pixelIndex}); + intr.areaLight, intr.p(), intr.n, intr.uv, intr.wo, r.lambda, + r.T_hat, r.uniPathPDF, r.lightPathPDF, r.prevIntrCtx, + (int)r.isSpecularBounce, r.pixelIndex}); } FloatTextureHandle displacement = material.GetDisplacement(); @@ -229,9 +230,9 @@ static __forceinline__ __device__ void ProcessClosestIntersection( q->Push(MaterialEvalWorkItem{ ptr, intr.pi, intr.n, intr.shading.n, intr.shading.dpdu, intr.shading.dpdv, intr.shading.dndu, intr.shading.dndv, - intr.uv, r.lambda, r.anyNonSpecularBounces, - r.T_hat, r.uniPathPDF, intr.wo, intr.time, r.etaScale, - getPayload()->mediumInterface, r.pixelIndex}); + intr.uv, r.lambda, r.anyNonSpecularBounces, intr.wo, r.pixelIndex, + r.T_hat, r.uniPathPDF, getPayload()->mediumInterface, + r.etaScale, intr.time}); }; material.Dispatch(enqueue); diff --git a/src/pbrt/gpu/pathintegrator.cpp b/src/pbrt/gpu/pathintegrator.cpp index 4b607dd1c5212466cb43ab0579313d7c56abef15..6c59f369ae3034905f79e556d64fbf390faca509 100644 --- a/src/pbrt/gpu/pathintegrator.cpp +++ b/src/pbrt/gpu/pathintegrator.cpp @@ -476,7 +476,7 @@ void GPUPathIntegrator::HandleRayFoundEmission(int depth) { "Handle emitters hit by indirect rays", hitAreaLightQueue, maxQueueSize, PBRT_GPU_LAMBDA(const HitAreaLightWorkItem w, int index) { // Find emitted radiance from surface that ray hit - SampledSpectrum Le = areaLight.L(w.p, w.n, w.uv, w.wo, w.lambda); + SampledSpectrum Le = w.areaLight.L(w.p, w.n, w.uv, w.wo, w.lambda); if (!Le) return; PBRT_DBG("Got Le %f %f %f %f from hit area light at depth %d\n", Le[0], Le[1], diff --git a/src/pbrt/gpu/workitems.h b/src/pbrt/gpu/workitems.h index 93fb2f4c1b6de8f57c2222d3cb471d92ae7fe95d..98f1dc4ed45e4c213d9f6986f1aa1d0810c40d9c 100644 --- a/src/pbrt/gpu/workitems.h +++ b/src/pbrt/gpu/workitems.h @@ -291,7 +291,6 @@ struct MaterialEvalWorkItem { int pixelIndex; SampledSpectrum T_hat, uniPathPDF; MediumInterface mediumInterface; - Float time; Float etaScale; Float time; };