diff --git a/atlas/graphics/deferred.cpp b/atlas/graphics/deferred.cpp index 3aada700..d868829d 100644 --- a/atlas/graphics/deferred.cpp +++ b/atlas/graphics/deferred.cpp @@ -254,6 +254,7 @@ buildGPUAreaLights(const std::vector &lights, int maxCount) { #ifdef METAL void Window::enableGlobalIllumination() { usesGlobalIllumination = true; + useSSR = true; ddgiSystem = std::make_shared(); ddgiSystem->sampleNormalMaps = false; ddgiSystem->init(); @@ -627,7 +628,9 @@ void Window::deferredRendering( if (usesGlobalIllumination && ddgiSystem != nullptr && ddgiSystem->probeSpace != nullptr && ddgiSystem->irradianceMap != nullptr && - ddgiSystem->irradianceMap->texture != nullptr) { + ddgiSystem->irradianceMap->texture != nullptr && + ddgiSystem->distanceMap != nullptr && + ddgiSystem->distanceMap->texture != nullptr) { lightPipeline->setUniform3f("ps.origin", ddgiSystem->probeSpace->originWorldSpace.x, ddgiSystem->probeSpace->originWorldSpace.y, @@ -653,12 +656,10 @@ void Window::deferredRendering( std::shared_ptr ddgiIrradianceTexture = ddgiSystem->irradianceMap->texture; - if (ddgiSystem->irradianceMapPrev != nullptr && - ddgiSystem->irradianceMapPrev->texture != nullptr && - ddgiSystem->frameIndex <= 1) { - ddgiIrradianceTexture = ddgiSystem->irradianceMapPrev->texture; - } + std::shared_ptr ddgiDistanceTexture = + ddgiSystem->distanceMap->texture; lightPipeline->bindTexture("irradianceMap", ddgiIrradianceTexture, 16); + lightPipeline->bindTexture("ddgiDistanceMap", ddgiDistanceTexture, 17); } else { lightPipeline->setUniform3f("ps.origin", 0.0f, 0.0f, 0.0f); lightPipeline->setUniform3f("ps.spacing", 1.0f, 1.0f, 1.0f); @@ -667,6 +668,8 @@ void Window::deferredRendering( lightPipeline->setUniform4f("ps.atlasParams", 0.0f, 0.0f, 0.0f, 0.0f); lightPipeline->bindTexture2D("irradianceMap", fallbackIrradianceTexture->textureID, 16); + lightPipeline->bindTexture2D("ddgiDistanceMap", + fallbackIrradianceTexture->textureID, 17); } #endif @@ -972,8 +975,6 @@ void Window::deferredRendering( lightPipeline->bindTextureCubemap( "skybox", fallbackSkyboxTexture->textureID, boundTextures); } - lightPipeline->setUniformBool( - "useIBL", scene->skybox != nullptr && scene->skybox->cubemap.id != 0); boundTextures++; lightPipeline->setUniform1f( diff --git a/include/atlas/core/default_shaders.h b/include/atlas/core/default_shaders.h index f7056622..9a48a9ca 100644 --- a/include/atlas/core/default_shaders.h +++ b/include/atlas/core/default_shaders.h @@ -109,7 +109,9 @@ static const AtlasPackedShaderSource COLOR_VERT = {COLOR_VERT_PARTS, 1}; static const char* const DDGI_PARTS[] = { R"(#include +#include using namespace metal; +using namespace raytracing; struct ProbeSpace { float3 origin; @@ -136,6 +138,8 @@ struct RaytracingSettings { uint probeUpdateOffset; uint probeUpdateStride; uint probeUpdateCount; + float3 skyColor; + uint useSkybox; }; struct Material { @@ -289,7 +293,8 @@ static inline bool rayTriangleMT(float3 ro, float3 rd, float3 v0, float3 v1, } static inline Hit traceScene(float3 ro, float3 rd, device const Triangle *tris, - uint triCount) { + instance_acceleration_structure sceneAS, + float maxDistance) { Hit best; best.t = INFINITY; best.hit = 0u; @@ -301,35 +306,36 @@ static inline Hit traceScene(float3 ro, float3 rd, device const Triangle *tris, best.triIndex = 0u; best._pad0 = 0u; - for (uint i = 0; i < triCount; i++) { - float t, u, v; - if (rayTriangleMT(ro, rd, tris[i].v0.xyz, tris[i].v1.xyz, - tris[i].v2.xyz, t, u, v)) { - if (t < best.t) { - best.t = t; - float w = 1.0f - u - v; - float3 n = normalize(tris[i].n0.xyz * w + tris[i].n1.xyz * u + - tris[i].n2.xyz * v); - best.n = n; - best.uv = - tris[i].uv0.xy * w + tris[i].uv1.xy * u + tris[i].uv2.xy * v; - float3 tRaw = tris[i].t0.xyz * w + tris[i].t1.xyz * u + - tris[i].t2.xyz * v; - float tLen2 = dot(tRaw, tRaw); - best.tangent = (tLen2 > 1e-10f) - ? tRaw * rsqrt(tLen2) - : float3(1.0f, 0.0f, 0.0f); - float3 bRaw = tris[i].b0.xyz * w + tris[i].b1.xyz * u + - tris[i].b2.xyz * v; - float bLen2 = dot(bRaw, bRaw); - best.bitangent = (bLen2 > 1e-10f) - ? bRaw * rsqrt(bLen2) - : float3(0.0f, 0.0f, 1.0f); - best.materialID = tris[i].materialID; - best.triIndex = i; - best.hit = 1u; - } - } + intersector isect; + isect.assume_geometry_type(geometry_type::triangle); + isect.set_triangle_cull_mode(triangle_cull_mode::none); + ray query; + query.origin = ro; + query.direction = rd; + query.min_distance = 0.0001f; + query.max_distance = max(maxDistance, query.min_distance); + auto intersection = isect.intersect(query, sceneAS, 0xFF); + if (intersection.type != intersection_type::none) { + uint i = intersection.primitive_id; + float2 bary = intersection.triangle_barycentric_coord; + float u = bary.x; + float v = bary.y; + float w = 1.0f - u - v; + best.t = intersection.distance; + best.n = normalize(tris[i].n0.xyz * w + tris[i].n1.xyz * u + + tris[i].n2.xyz * v); + best.uv = tris[i].uv0.xy * w + tris[i].uv1.xy * u + tris[i].uv2.xy * v; + float3 tRaw = tris[i].t0.xyz * w + tris[i].t1.xyz * u + tris[i].t2.xyz * v; + float tLen2 = dot(tRaw, tRaw); + best.tangent = (tLen2 > 1e-10f) ? tRaw * rsqrt(tLen2) + : float3(1.0f, 0.0f, 0.0f); + float3 bRaw = tris[i].b0.xyz * w + tris[i].b1.xyz * u + tris[i].b2.xyz * v; + float bLen2 = dot(bRaw, bRaw); + best.bitangent = (bLen2 > 1e-10f) ? bRaw * rsqrt(bLen2) + : float3(0.0f, 0.0f, 1.0f); + best.materialID = tris[i].materialID; + best.triIndex = i; + best.hit = 1u; } return best; @@ -382,10 +388,23 @@ static inline float4 sampleMaterialTexture( texture2d materialTexture17, texture2d materialTexture18, texture2d materialTexture19, texture2d materialTexture20, texture2d materialTexture21, texture2d materialTexture22, - texture2d materialTexture23) { + texture2d materialTexture23, texture2d materialTexture24, + texture2d materialTexture25, texture2d materialTexture26, + texture2d materialTexture27, texture2d materialTexture28, + texture2d materialTexture29, texture2d materialTexture30, + texture2d materialTexture31, texture2d materialTexture32, + texture2d materialTexture33, texture2d materialTexture34, + texture2d materialTexture35, texture2d materialTexture36, + texture2d materialTexture37, texture2d materialTexture38, + texture2d materialTexture39, texture2d materialTexture40, + texture2d materialTexture41, texture2d materialTexture42, + texture2d materialTexture43, texture2d materialTexture44, + texture2d materialTexture45, texture2d materialTexture46, + texture2d materialTexture47) { switch (textureIndex) { case 0: - return materialTexture0.sample(materialTexSampler, uv); + return materialTexture0.sample)", +R"((materialTexSampler, uv); case 1: return materialTexture1.sample(materialTexSampler, uv); case 2: @@ -416,8 +435,7 @@ static inline float4 sampleMaterialTexture( return materialTexture14.sample(materialTexSampler, uv); case 15: return materialTexture15.sample(materialTexSampler, uv); - case)", -R"( 16: + case 16: return materialTexture16.sample(materialTexSampler, uv); case 17: return materialTexture17.sample(materialTexSampler, uv); @@ -433,6 +451,54 @@ R"( 16: return materialTexture22.sample(materialTexSampler, uv); case 23: return materialTexture23.sample(materialTexSampler, uv); + case 24: + return materialTexture24.sample(materialTexSampler, uv); + case 25: + return materialTexture25.sample(materialTexSampler, uv); + case 26: + return materialTexture26.sample(materialTexSampler, uv); + case 27: + return materialTexture27.sample(materialTexSampler, uv); + case 28: + return materialTexture28.sample(materialTexSampler, uv); + case 29: + return materialTexture29.sample(materialTexSampler, uv); + case 30: + return materialTexture30.sample(materialTexSampler, uv); + case 31: + return materialTexture31.sample(materialTexSampler, uv); + case 32: + return materialTexture32.sample(materialTexSampler, uv); + case 33: + return materialTexture33.sample(materialTexSampler, uv); + case 34: + return materialTexture34.sample(materialTexSampler, uv); + case 35: + return materialTexture35.sample(materialTexSampler, uv); + case 36: + return materialTexture36.sample(materialTexSampler, uv); + case 37: + return materialTexture37.sample(materialTexSampler, uv); + case 38: + return materialTexture38.sample(materialTexSampler, uv); + case 39: + return materialTexture39.sample(materialTexSampler, uv); + case 40: + return materialTexture40.sample(materialTexSampler, uv); + case 41: + return materialTexture41.sample(materialTexSampler, uv); + case 42: + return materialTexture42.sample(materialTexSampler, uv); + case 43: + return materialTexture43.sample(materialTexSampler, uv); + case 44: + return materialTexture44.sample(materialTexSampler, uv); + case 45: + return materialTexture45.sample(materialTexSampler, uv); + case 46: + return materialTexture46.sample(materialTexSampler, uv); + case 47: + return materialTexture47.sample(materialTexSampler, uv); default: break; } @@ -453,7 +519,19 @@ static inline void resolveMaterialParameters( texture2d materialTexture17, texture2d materialTexture18, texture2d materialTexture19, texture2d materialTexture20, texture2d materialTexture21, texture2d materialTexture22, - texture2d materialTexture23, thread float3 &albedo, + texture2d materialTexture23, texture2d materialTexture24, + texture2d materialTexture25, texture2d materialTexture26, + texture2d materialTexture27, texture2d materialTexture28, + texture2d materialTexture29, texture2d materialTexture30, + texture2d materialTexture31, texture2d materialTexture32, + texture2d materialTexture33, texture2d materialTexture34, + texture2d materialTexture35, texture2d materialTexture36, + texture2d materialTexture37, texture2d materialTexture38, + texture2d materialTexture39, texture2d materialTexture40, + texture2d materialTexture41, texture2d materialTexture42, + texture2d materialTexture43, texture2d materialTexture44, + texture2d materialTexture45, texture2d materialTexture46, + texture2d materialTexture47, thread float3 &albedo, thread float &metallic, thread float &roughness, thread float &ao, thread float3 &emissive, thread int &normalTextureIndex, thread float &normalStrength) { @@ -491,7 +569,20 @@ static inline void resolveMaterialParameters( materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, - materialTexture22, materialTexture23) + materialTexture22, materialTexture23, + materialTexture24, materialTexture25, + materialTexture26, materialTexture27, + materialTexture28, materialTexture29, + materialTexture30, materialTexture31, + materialTexture32, materialTexture33, + materialTexture34, materialTexture35, + materialTexture36, materialTexture37, + materialTexture38, materialTexture39, + materialTexture40, materialTexture41, + materialTexture42)", +R"(, materialTexture43, + materialTexture44, materialTexture45, + materialTexture46, materialTexture47) .xyz, float3(0.0f), float3(1.0f)); } @@ -506,7 +597,15 @@ static inline void resolveMaterialParameters( materialTexture14, materialTexture15, materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, materialTexture22, - materialTexture23); + materialTexture23, materialTexture24, materialTexture25, + materialTexture26, materialTexture27, materialTexture28, + materialTexture29, materialTexture30, materialTexture31, + materialTexture32, materialTexture33, materialTexture34, + materialTexture35, materialTexture36, materialTexture37, + materialTexture38, materialTexture39, materialTexture40, + materialTexture41, materialTexture42, materialTexture43, + materialTexture44, materialTexture45, materialTexture46, + materialTexture47); float metallicValue = metallicSample.x; if (mat.roughnessTextureIndex == mat.metallicTextureIndex) { metallicValue = metallicSample.z; @@ -524,7 +623,15 @@ static inline void resolveMaterialParameters( materialTexture14, materialTexture15, materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, materialTexture22, - materialTexture23); + materialTexture23, materialTexture24, materialTexture25, + materialTexture26, materialTexture27, materialTexture28, + materialTexture29, materialTexture30, materialTexture31, + materialTexture32, materialTexture33, materialTexture34, + materialTexture35, materialTexture36, materialTexture37, + materialTexture38, materialTexture39, materialTexture40, + materialTexture41, materialTexture42, materialTexture43, + materialTexture44, materialTexture45, materialTexture46, + materialTexture47); float roughnessValue = roughnessSample.x; if (mat.roughnessTextureIndex == mat.metallicTextureIndex) { roughnessValue = roughnessSample.y; @@ -543,7 +650,19 @@ static inline void resolveMaterialParameters( materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, - materialTexture22, materialTexture23) + materialTexture22, materialTexture23, + materialTexture24, materialTexture25, + materialTexture26, materialTexture27, + materialTexture28, materialTexture29, + materialTexture30, materialTexture31, + materialTexture32, materialTexture33, + materialTexture34, materialTexture35, + materialTexture36, materialTexture37, + materialTexture38, materialTexture39, + materialTexture40, materialTexture41, + materialTexture42, materialTexture43, + materialTexture44, materialTexture45, + materialTexture46, materialTexture47) .x, 0.0f, 1.0f); } @@ -568,14 +687,25 @@ static inline float3 resolveNormal( texture2d materialTexture17, texture2d materialTexture18, texture2d materialTexture19, texture2d materialTexture20, texture2d materialTexture21, texture2d materialTexture22, - texture2d materialTexture23) { + texture2d materialTexture23, texture2d materialTexture24, + texture2d materialTexture25, texture2d materialTexture26, + texture2d materialTexture27, texture2d materialTexture28, + texture2d materialTexture29, texture2d materialTexture30, + texture2d materialTexture31, texture2d materialTexture32, + texture2d materialTexture33, texture2d materialTexture34, + texture2d materialTexture35, texture2d materialTexture36, + texture2d materialTexture37, texture2d materialTexture38, + texture2d materialTexture39, texture2d materialTexture40, + texture2d materialTexture41, texture2d materialTexture42, + texture2d materialTexture43, texture2d materialTexture44, + texture2d materialTexture45, texture2d materialTexture46, + texture2d materialTexture47) { float3 N = safeNormalize(n, float3(0.0f, 1.0f, 0.0f)); float3 T = safeNormalize(tangent - N * dot(N, tangent), float3(1.0f, 0.0f, 0.0f)); float3 B = safeNormalize(bitangent - N * dot(N, bitangent), cross(N, T)); if (dot(cross(T, B), cross(T, B)) <= 1e-10f) { float3 up = (fabs(N.y) < 0.999f) ? float3(0.0f, 1.0f, 0.0f) - )", -R"( : float3(1.0f, 0.0f, 0.0f); + : float3(1.0f, 0.0f, 0.0f); T = safeNormalize(cross(up, N), float3(1.0f, 0.0f, 0.0f)); B = safeNormalize(cross(N, T), float3(0.0f, 0.0f, 1.0f)); } @@ -586,13 +716,26 @@ R"( : float3(1.0f, 0.0f, 0.0f); materialTexture1, materialTexture2, materialTexture3, materialTexture4, materialTexture5, materialTexture6, materialTexture7, materialTexture8, materialTexture9, - materialTexture10, materialTexture11, + materialTexture10, m)", +R"(aterialTexture11, materialTexture12, materialTexture13, materialTexture14, materialTexture15, materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, - materialTexture22, materialTexture23) + materialTexture22, materialTexture23, + materialTexture24, materialTexture25, + materialTexture26, materialTexture27, + materialTexture28, materialTexture29, + materialTexture30, materialTexture31, + materialTexture32, materialTexture33, + materialTexture34, materialTexture35, + materialTexture36, materialTexture37, + materialTexture38, materialTexture39, + materialTexture40, materialTexture41, + materialTexture42, materialTexture43, + materialTexture44, materialTexture45, + materialTexture46, materialTexture47) .xyz; texN = texN * 2.0f - 1.0f; texN.xy *= normalStrength; @@ -602,20 +745,66 @@ R"( : float3(1.0f, 0.0f, 0.0f); return N; } -static inline float3 sampleSky(float3 d) { +static inline float2 octEncode(float3 n) { + n /= max(fabs(n.x) + fabs(n.y) + fabs(n.z), 1e-6f); + float2 e = n.xy; + if (n.z < 0.0f) { + float2 signNotZero = + float2(e.x >= 0.0f ? 1.0f : -1.0f, e.y >= 0.0f ? 1.0f : -1.0f); + e = (1.0f - fabs(e.yx)) * signNotZero; + } + return e; +} + +constexpr sampler ddgiLinearSampler(coord::normalized, address::clamp_to_edge, + filter::linear); + +static inline float3 samplePreviousIrradiance(texture2d previous, + constant ProbeSpace &ps, + float3 posWS, float3 direction) { + uint3 counts = uint3(ps.probeCount); + if (counts.x == 0u || counts.y == 0u || counts.z == 0u) { + return float3(0.0f); + } + float3 grid = clamp((posWS - ps.origin) / max(ps.spacing, float3(1e-4f)), + float3(0.0f), float3(counts - 1u)); + uint3 coord = uint3(round(grid)); + uint probeIndex = coord.x + counts.x * (coord.y + counts.y * coord.z); + uint border = uint(ps.atlasParams.x); + uint innerRes = uint(ps.atlasParams.y); + uint probesPerRow = uint(ps.atlasParams.z); + uint tileRes = innerRes + 2u * border; + uint2 tile = uint2(probeIndex % probesPerRow, probeIndex / probesPerRow); + float2 inner = octEncode(safeNormalize(direction, float3(0.0f, 1.0f, 0.0f))) * + 0.5f + + 0.5f; + float2 pixel = float2(tile * tileRes + border) + + inner * float(max(innerRes, 1u) - 1u) + 0.5f; + float3 history = previous.sample( + ddgiLinearSampler, + pixel / float2(previous.get_width(), + previous.get_height())) + .xyz; + return all(isfinite(history)) ? max(history, float3(0.0f)) + : float3(0.0f); +} + +static inline float3 sampleSky(float3 d, texturecube skybox, + float3 skyColor, uint useSkybox) { + if (useSkybox != 0u) { + return max(skybox.sample(ddgiLinearSampler, d).xyz, float3(0.0f)); + } float t = clamp(d.y * 0.5f + 0.5f, 0.0f, 1.0f); - return mix(float3(0.02f, 0.023f, 0.028f), - float3(0.12f, 0.14f, 0.18f), - t); + return mix(skyColor * 0.08f, skyColor, t); } static inline float shadowVisibility(float3 ro, float3 rd, float maxT, device const Triangle *tris, - uint triCount) { + instance_acceleration_structure sceneAS) { if (maxT <= 1e-4f) { return 1.0f; } - Hit h = traceScene(ro, rd, tris, triCount); + Hit h = traceScene(ro, rd, tris, sceneAS, maxT); if (!h.hit) { return 1.0f; } @@ -632,7 +821,7 @@ static inline float giNdotL(float3 n, float3 l) { static inline float3 evaluateDirectLights( float3 posWS, float3 normalWS, float bias, float maxDistance, - device const Triangle *tris, uint triCount, + device const Triangle *tris, instance_acceleration_structure sceneAS, device const DirectionalLight *directionalLights, uint directionalLightCount, device const PointLight *pointLights, uint pointLightCount, device const SpotLight *spotLights, uint spotLightCount, @@ -647,8 +836,10 @@ static inline float3 evaluateDirectLights( if (ndl <= 0.0f) continue; + float visibility = shadowVisibility(posWS + n * bias, L, maxDistance, + tris, sceneAS); sum += directionalLights[i].diffuse * - max(0.0f, directionalLights[i].intensity) * ndl; + max(0.0f, directionalLights[i].intensity) * ndl * visibility; } for (uint i = 0; i < pointLightCount; i++) { @@ -668,8 +859,10 @@ static inline float3 evaluateDirectLights( pointLights[i].quadratic * dist * dist, 1e-4f); float fade = 1.0f - smoothstep(radius * 0.9f, radius, dist); + float visibility = shadowVisibility(posWS + n * bias, L, dist - bias, + tris, sceneAS); sum += pointLights[i].diffuse * max(0.0f, pointLights[i].intensity) * - attenuation * fade * ndl; + attenuation * fade * ndl * visibility; } for (uint i = 0; i < spotLightCount; i++) { @@ -697,8 +890,10 @@ static inline float3 evaluateDirectLights( float attenuation = 1.0f / ((1.0f + (dist / range)) + ((dist * dist) / (range * range))); float fade = 1.0f - smoothstep(range * 0.9f, range, dist); + float visibility = shadowVisibility(posWS + n * bias, L, dist - bias, + tris, sceneAS); sum += spotLights[i].diffuse * max(0.0f, spotLights[i].intensity) * cone * - attenuation * fade * ndl; + attenuation * fade * ndl * visibility; } for (uint i = 0; i < areaLightCount; i++) { @@ -710,7 +905,8 @@ static inline float3 evaluateDirectLights( float3 toPoint = posWS - center; float s = clamp(dot(toPoint, right), -halfSize.x, halfSize.x); float t = clamp(dot(toPoint, up), -halfSize.y, halfSize.y); - float3 closest = center + right * s + up * t; + float3 closest = center + right * s + up *)", +R"( t; float3 Lvec = closest - posWS; float dist = length(Lvec); @@ -734,8 +930,10 @@ static inline float3 evaluateDirectLights( float attenuation = 1.0f / ((1.0f + (dist / range)) + ((dist * dist) / (range * range))); float fade = 1.0f - smoothstep(range * 0.9f, range, dist); + float visibility = shadowVisibility(posWS + n * bias, L, dist - bias, + tris, sceneAS); sum += areaLights[i].diffuse * max(0.0f, areaLights[i].intensity) * - facing * attenuation * fade * ndl; + facing * attenuation * fade * ndl * visibility; } return sum; @@ -778,11 +976,11 @@ kernel void main0(device float4 *probeRadianceOut [[buffer(0)]], constant SceneCounts &sc [[buffer(3)]], constant ProbeSpace &ps [[buffer(4)]], constant RaytracingSettings &rt [[buffer(5)]], - device const DirectionalLight *d)", -R"(irectionalLights [[buffer(6)]], + device const DirectionalLight *directionalLights [[buffer(6)]], device const PointLight *pointLights [[buffer(7)]], device const SpotLight *spotLights [[buffer(8)]], device const AreaLight *areaLights [[buffer(9)]], + instance_acceleration_structure sceneAS [[buffer(10)]], texture2d materialTexture0 [[texture(10)]], texture2d materialTexture1 [[texture(11)]], texture2d materialTexture2 [[texture(12)]], @@ -807,6 +1005,32 @@ R"(irectionalLights [[buffer(6)]], texture2d materialTexture21 [[texture(31)]], texture2d materialTexture22 [[texture(32)]], texture2d materialTexture23 [[texture(33)]], + texture2d materialTexture24 [[texture(34)]], + texture2d materialTexture25 [[texture(35)]], + texture2d materialTexture26 [[texture(36)]], + texture2d materialTexture27 [[texture(37)]], + texture2d materialTexture28 [[texture(38)]], + texture2d materialTexture29 [[texture(39)]], + texture2d materialTexture30 [[texture(40)]], + texture2d materialTexture31 [[texture(41)]], + texture2d materialTexture32 [[texture(42)]], + texture2d materialTexture33 [[texture(43)]], + texture2d materialTexture34 [[texture(44)]], + texture2d materialTexture35 [[texture(45)]], + texture2d materialTexture36 [[texture(46)]], + texture2d materialTexture37 [[texture(47)]], + texture2d materialTexture38 [[texture(48)]], + texture2d materialTexture39 [[texture(49)]], + texture2d materialTexture40 [[texture(50)]], + texture2d materialTexture41 [[texture(51)]], + texture2d materialTexture42 [[texture(52)]], + texture2d materialTexture43 [[texture(53)]], + texture2d materialTexture44 [[texture(54)]], + texture2d materialTexture45 [[texture(55)]], + texture2d materialTexture46 [[texture(56)]], + texture2d materialTexture47 [[texture(57)]], + texturecube skybox [[texture(60)]], + texture2d previousIrradiance [[texture(61)]], uint tid [[thread_position_in_grid]]) { const float PI = 3.14159265359f; uint totalProbes = (uint)ps.atlasParams.w; @@ -848,8 +1072,9 @@ R"(irectionalLights [[buffer(6)]], float3 ro = probePos + rayDir * bias; Hit h; float selfHitThreshold = bias * 4.0f; - for (uint escapeStep = 0u; escapeStep < 1u; escapeStep++) { - h = traceScene(ro, rayDir, tris, sc.triCount); + for (uint escapeStep = 0u)", +R"(; escapeStep < 1u; escapeStep++) { + h = traceScene(ro, rayDir, tris, sceneAS, maxDistance); if (h.hit == 0u || h.t >= selfHitThreshold) { break; } @@ -862,7 +1087,7 @@ R"(irectionalLights [[buffer(6)]], float3 radiance = float3(0.0f); if (h.hit == 0u || h.t > maxDistance) { - radiance = sampleSky(rayDir); + radiance = sampleSky(rayDir, skybox, rt.skyColor, rt.useSkybox); } else { float3 hitPos = ro + rayDir * h.t; float3 albedo; @@ -882,7 +1107,15 @@ R"(irectionalLights [[buffer(6)]], materialTexture13, materialTexture14, materialTexture15, materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, - materialTexture22, materialTexture23, albedo, metallic, roughness, + materialTexture22, materialTexture23, materialTexture24, + materialTexture25, materialTexture26, materialTexture27, + materialTexture28, materialTexture29, materialTexture30, + materialTexture31, materialTexture32, materialTexture33, + materialTexture34, materialTexture35, materialTexture36, + materialTexture37, materialTexture38, materialTexture39, + materialTexture40, materialTexture41, materialTexture42, + materialTexture43, materialTexture44, materialTexture45, + materialTexture46, materialTexture47, albedo, metallic, roughness, ao, emissive, normalTextureIndex, normalStrength); float3 hitNormal = resolveNormal( @@ -895,21 +1128,35 @@ R"(irectionalLights [[buffer(6)]], materialTexture14, materialTexture15, materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, materialTexture22, - materialTexture23); + materialTexture23, materialTexture24, materialTexture25, + materialTexture26, materialTexture27, materialTexture28, + materialTexture29, materialTexture30, materialTexture31, + materialTexture32, materialTexture33, materialTexture34, + materialTexture35, materialTexture36, materialTexture37, + materialTexture38, materialTexture39, materialTexture40, + materialTexture41, materialTexture42, materialTexture43, + materialTexture44, materialTexture45, materialTexture46, + materialTexture47); if (dot(hitNormal, -rayDir) < 0.0f) { hitNormal = -hitNormal; } float3 direct = evaluateDirectLights( - hitPos, hitNormal, bias, maxDistance, tris, sc.triCount, + hitPos, hitNormal, bias, maxDistance, tris, sceneAS, directionalLights, sc.directionalLightCount, pointLights, sc.pointLightCount, spotLights, sc.spotLightCount, areaLights, sc.areaLightCount); - float diffuseWeight = - (1.0f - metallic) * mix(0.35f, 1.0f, 1.0f - roughness); - radiance = direct * albedo * diffuseWeight * max(ao, 0.05f) + emissive; + float diffuseWeight = 1.0f - metallic; + float3 diffuseResponse = albedo * diffuseWeight * max(ao, 0.05f) / PI; + float3 previousBounce = + rt.frameIndex >= max(rt.probeUpdateStride, 1u) + ? samplePreviousIrradiance(previousIrradiance, ps, + hitPos + hitNormal * bias, hitNormal) + : float3(0.0f); + float3 indirect = previousBounce * diffuseResponse * 0.35f; + radiance = direct * diffuseResponse + indirect + emissive; radiance = clamp(radiance, float3(0.0f), float3(16.0f)); } @@ -918,7 +1165,7 @@ R"(irectionalLights [[buffer(6)]], } )", }; -static const AtlasPackedShaderSource DDGI = {DDGI_PARTS, 4}; +static const AtlasPackedShaderSource DDGI = {DDGI_PARTS, 6}; static const char* const DDGI_WRITE_PARTS[] = { R"(#include @@ -987,6 +1234,8 @@ static inline float3 sphericalFibonacci(uint index, uint count, uint frameIndex) kernel void main0(texture2d outTexture [[texture(0)]], texture2d prevTexture [[texture(1)]], + texture2d outDistance [[texture(2)]], + texture2d prevDistance [[texture(3)]], device float4 *probeRadiance [[buffer(0)]], constant ProbeSpace &ps [[buffer(1)]], constant RaytracingSettings &rt [[buffer(2)]], @@ -1005,6 +1254,7 @@ kernel void main0(texture2d outTexture [[texture(0)]], if (tileRes == 0u || probesPerRow == 0u || totalProbes == 0u || innerRes == 0u) { outTexture.write(prevTexture.read(gid), gid); + outDistance.write(prevDistance.read(gid), gid); return; } @@ -1014,6 +1264,7 @@ kernel void main0(texture2d outTexture [[texture(0)]], if (probeIndex >= totalProbes) { outTexture.write(prevTexture.read(gid), gid); + outDistance.write(prevDistance.read(gid), gid); return; } @@ -1026,6 +1277,7 @@ kernel void main0(texture2d outTexture [[texture(0)]], (((probeIndex - updateOffset) % updateStride) == 0u)); if (!probeIsActive) { outTexture.write(prevTexture.read(gid), gid); + outDistance.write(prevDistance.read(gid), gid); return; } @@ -1045,12 +1297,13 @@ kernel void main0(texture2d outTexture [[texture(0)]], float3 sum = float3(0.0f); float weightSum = 0.0f; + float distanceSum = 0.0f; + float distanceSquaredSum = 0.0f; float nearHitCount = 0.0f; - float missCount = 0.0f; float spacingScale = max(max(ps.spacing.x, max(ps.spacing.y, ps.spacing.z)), 1e-4f); float nearHitThreshold = - max(max(rt.normalBias * 1.2f, spacingScale * 0.015f), 0.0008f); + max(max(rt.normalBias * 2.0f, spacingScale * 0.1f), 0.002f); for (uint r = 0; r < raysPerProbe; r += rayStep) { sampledRayCount++; @@ -1058,8 +1311,6 @@ kernel void main0(texture2d outTexture [[texture(0)]], float hitDistance = raySample.w; if (hitDistance > 0.0f && hitDistance < nearHitThreshold) { nearHitCount += 1.0f; - } else if (hitDistance <= 0.0f) { - missCount += 1.0f; } float3 rayDir = sphericalFibonacci(r, raysPerProbe, rt.frameIndex); @@ -1067,19 +1318,24 @@ kernel void main0(texture2d outTexture [[texture(0)]], if (w > 1e-6f) { float3 rad = raySample.xyz; if (all(isfinite(rad))) { - float lum = dot(rad, float3(0.2126f, 0.7152f, 0.0722f)); - float compression = 1.0f / (1.0f + lum * 0.25f); - rad *= compression; sum += rad * w; weightSum += w; + float distance = hitDistance > 0.0f + ? min(hitDistance, rt.maxRayDistance) + : rt.maxRayDistance; + distanceSum += distance * w; + distanceSquaredSum += distance * distance * w; } } } float3 irradiance = float3(0.0f); + float2 distanceMoments = float2(rt.maxRayDistance, + rt.maxRayDistance * rt.maxRayDistance); float invRayCount = 1.0f / float(max(sampledRayCount, 1u)); if (weightSum > 1e-6f) { irradiance = sum * (FOUR_PI * invRayCount); + distanceMoments = float2(distanceSum, distanceSquaredSum) / weightSum; } if (!all(isfinite(irradiance))) { @@ -1087,24 +1343,30 @@ kernel void main0(texture2d outTexture [[texture(0)]], } float4 prev = prevTexture.read(gid); + float4 previousDistance = prevDistance.read(gid); float3 prevValue = all(isfinite(prev.xyz)) ? prev.xyz : float3(0.0f); float prevValidity = isfinite(prev.w) ? clamp(prev.w, 0.0f, 1.0f) : 1.0f; float nearFraction = nearHitCount * invRayCount; - float missFraction = missCount * invRayCount; float nearPenalty = smoothstep(0.82f, 0.995f, nearFraction); - float missPenalty = smoothstep(0.95f, 1.0f, missFraction); - float probeValidity = (1.0f - nearPenalty) * (1.0f - missPenalty); - probeValidity = clamp(probeValidity, 0.005f, 1.0f); + float probeValidity = 1.0f - nearPenalty; + probeValidity = clamp(probeValidity, 0.0f, 1.0f); float h = clamp(rt.hysteresis, 0.0f, 0.995f); - float3 blended = (rt.frameIndex == 0u) ? irradiance : mix(irradiance, prevValue, h); + bool firstProbeUpdate = rt.frameIndex < updateStride; + float3 blended = firstProbeUpdate ? irradiance : mix(irradiance, prevValue, h); + float2 blendedDistance = + firstProbeUpdate + ? distanceMoments + : mix(distanceMoments, previousDistance.xy, h); float blendedValidity = - (rt.frameIndex == 0u) + firstProbeUpdate ? probeValidity : mix(probeValidity, prevValidity, h); outTexture.write(float4(max(blended, float3(0.0f)), blendedValidity), gid); + outDistance.write(float4(max(blendedDistance, float2(0.0f)), + blendedValidity, 0.0f), gid); } )", }; @@ -4196,6 +4458,7 @@ sampleProbeDirectionalRadiance(texture2d ddgiTexture, static inli)", R"(ne float3 sampleDDGIIrradiance(texture2d ddgiTexture, + texture2d ddgiDistance, constant ProbeSpace &ps, float3 posWS, float3 normalWS) { uint3 counts = uint3((uint)ps.probeCount.x, (uint)ps.probeCount.y, @@ -4232,13 +4495,27 @@ R"(ne float3 sampleDDGIIrradiance(texture2d ddgiTexture, uint pIndex = probeIndexFromCoord(nearest, counts); float4 nearestSample = sampleProbeDirectionalRadiance( ddgiTexture, ps, pIndex, atlasW, atlasH, safeNormal); + float3 nearestProbePos = ps.origin + float3(nearest) * safeSpacing; + float3 nearestDirection = safeNormalizeDDGI( + posWS - nearestProbePos, safeNormal); + float4 nearestMoments = sampleProbeDirectionalRadiance( + ddgiDistance, ps, pIndex, atlasW, atlasH, nearestDirection); + float nearestDistance = length(posWS - nearestProbePos); + float nearestVariance = max(nearestMoments.y - nearestMoments.x * + nearestMoments.x, + 0.0001f); + float nearestDelta = max(nearestDistance - nearestMoments.x, 0.0f); + float nearestVisibility = nearestDelta <= 0.0f + ? 1.0f + : nearestVariance / + (nearestVariance + nearestDelta * + nearestDelta); float nearestValidity = isfinite(nearestSample.w) ? clamp(nearestSample.w, 0.0f, 1.0f) : 0.0f; - nearestValidity = mix(0.05f, 1.0f, nearestValidity); float3 nearestIrr = all(isfinite(nearestSample.xyz)) ? nearestSample.xyz : float3(0.0f); - nearestIrr *= nearestValidity; + nearestIrr *= nearestValidity * nearestVisibility; return max(nearestIrr, float3(0.0f)); } @@ -4288,20 +4565,33 @@ R"(ne float3 sampleDDGIIrradiance(texture2d ddgiTexture, float4 irrSample = sampleProbeDirectionalRadiance( ddgiTexture, ps, pIndex, atlasW, atlasH, safeNormal); + float3 probeToSurface = -surfaceToProbe; + float4 distanceSample = sampleProbeDirectionalRadiance( + ddgiDistance, ps, pIndex, atlasW, atlasH, + safeNormalizeDDGI(probeToSurface, safeNormal)); + float variance = max(distanceSample.y - + distanceSample.x * distanceSample.x, + 0.0001f); + float delta = max(sDist - distanceSample.x, 0.0f); + float visibility = delta <= 0.0f + ? 1.0f + : variance / (variance + delta * delta); + visibility = visibility * visibility * visibility; float probeValidity = isfinite(irrSample.w) ? clamp(irrSample.w, 0.0f, 1.0f) : 0.0f; - float validityW = mix(0.05f, 1.0f, probeValidity); + float validityW = probeValidity; float3 irr = irrSample.xyz; if (!all(isfinite(irr))) { irr = float3(0.0f); validityW = 0.0f; } irr = max(irr, float3(0.0f)); - float weightedW = w * validityW; + float weightedW = w * validityW * max(visibility, 0.001f); result += irr * weightedW; weightSum += weightedW; - float noBackfaceW = trilinearW * distanceW * validityW; + float noBackfaceW = trilinearW * distanceW * validityW * + max(visibility, 0.001f); resultNoBackface += irr * noBackfaceW; weightSumNoBackface += noBackfaceW; } @@ -4324,7 +4614,6 @@ R"(ne float3 sampleDDGIIrradiance(texture2d ddgiTexture, ddgiTexture, ps, nearestIndex, atlasW, atlasH, safeNormal); float fallbackValidity = isfinite(fallbackSample.w) ? clamp(fallbackSample.w, 0.0f, 1.0f) : 0.0f; - fallbackValidity = mix(0.05f, 1.0f, fallbackValidity); float3 fallback = all(isfinite(fallbackSample.xyz)) ? fallbackSample.xyz : float3(0.0f); fallback *= fallbackValidity; @@ -4338,7 +4627,8 @@ fragment main0_out main0( main0_in in [[stage_in]], constant UBO &_526 [[buffer(0)]], constant Environment &environment [[buffer(1)]], constant PushConstants &_1355 [[buffer(2)]], - device ShadowParams &_1372 [[buffer(3)]], + device ShadowParams &_1372 [[b)", +R"(uffer(3)]], device DirectionalLights &_1422 [[buffer(4)]], device PointLights &_1465 [[buffer(5)]], device SpotLights &_1510 [[buffer(6)]], @@ -4362,6 +4652,7 @@ fragment main0_out main0( texture2d gMaterial [[texture(14)]], texture2d ssao [[texture(15)]], texture2d irradianceMap [[texture(16)]], + texture2d ddgiDistanceMap [[texture(17)]], sampler texture1Smplr [[sampler(0)]], sampler texture2Smplr [[sampler(1)]], sampler texture3Smplr [[sampler(2)]], sampler texture4Smplr [[sampler(3)]], sampler texture5Smplr [[sampler(4)]], sampler cubeMap1Smplr [[sampler(5)]], @@ -4371,8 +4662,7 @@ fragment main0_out main0( sampler gNormalSmplr [[sampler(12)]], sampler gAlbedoSpecSmplr [[sampler(13)]], sampler gMaterialSmplr [[sampler(14)]], sampler ssaoSmplr [[sampler(15)]]) { - main0_out ou)", -R"(t{}; + main0_out out{}; float4 gPositionSample = gPosition.sample(gPositionSmplr, in.TexCoord); float3 FragPos = gPositionSample.xyz; if (!all(isfinite(FragPos))) { @@ -4504,7 +4794,8 @@ R"(t{}; _1397.textureIndex = _1372.shadowParams[i].textureIndex; _1397.farPlane = _1372.shadowParams[i].farPlane; _1397._pad1 = _1372.shadowParams[i]._pad1; - _1397.lightPos = float3(_1372.shadowParams[i].lightPos); + _1397.lightPos = float3(_13)", +R"(72.shadowParams[i].lightPos); _1397.lightType = _1372.shadowParams[i].lightType; ShadowParameters param_2 = _1397; float3 param_3 = FragPos; @@ -4540,8 +4831,7 @@ R"(t{}; } directionalShadow = fast::clamp(directionalShadow * 0.85000002384185791015625, 0.0, 1.0); - spotShadow = fast::clamp(spotShadow * 0.8500000238418579)", -R"(1015625, 0.0, 1.0); + spotShadow = fast::clamp(spotShadow * 0.85000002384185791015625, 0.0, 1.0); areaShadow = fast::clamp(areaShadow * 0.449999988079071044921875, 0.0, 1.0); pointShadow = fast::clamp(pointShadow * 0.85000002384185791015625, 0.0, 1.0); @@ -4677,7 +4967,8 @@ R"(1015625, 0.0, 1.0); float3 param_40 = albedo; float param_41 = metallic; float param_42 = roughness; - float3 _1714 = getRimLight(param_36, param_37, param_38, param_39, param_40, + float3 _1714 = getRimLight(param_36, )", +R"(param_37, param_38, param_39, param_40, param_41, param_42, _526, environment); float3 rimResult = _1714; float3 lighting = @@ -4687,13 +4978,15 @@ R"(1015625, 0.0, 1.0); float3 ambientBase = ((ambientLight.color.xyz * ambientLight.intensity) * albedo) * occlusion; - float3 ambient = ambientBase; + bool ddgiEnabled = ps.atlasParams.w > 0.0f; + float3 ambient = ddgiEnabled ? ambientBase * 0.05f : ambientBase; float ddgiSampleBias = max(max(ps.spacing.x, max(ps.spacing.y, ps.spacing.z)) * 0.05f, 0.002f); float3 ddgiSamplePos = FragPos + ddgiNormal * ddgiSampleBias; float3 ddgiIrradiance = - sampleDDGIIrradiance(irradianceMap, ps, ddgiSamplePos, ddgiNormal); + sampleDDGIIrradiance(irradianceMap, ddgiDistanceMap, ps, ddgiSamplePos, + ddgiNormal); if (!all(isfinite(ddgiIrradiance))) { ddgiIrradiance = float3(0.0f); } @@ -4706,46 +4999,11 @@ R"(1015625, 0.0, 1.0); } const float INV_PI = 0.31830988618379067153776752674503; - float ddgiLuma = dot(ddgiIrradiance, float3(0.2126f, 0.7152f, 0.0722f)); - float3 ddgiChroma = ddgiIrradiance - float3(ddgiLuma); - float3 boostedIrradiance = - max(float3(ddgiLuma * 0.1500000059604644775390625) + - ddgiChroma * 1.35000002384185791015625, - float3(0.0f)); - float3 bleedAlbedo = albedo; - float3 ddgiDiffuse = boostedIrradiance * bleedAlbedo * INV_PI * - (1.0f - metallic) * ddgiGain * - 0.85000002384185791015625; - float sideFactor = clamp(1.0f - abs(N.y), 0.0f, 1.0f); - float ddgiSurfaceFactor = - 0.550000011920928955078125 + sideFactor * 0.44999998807907104492)", -R"(1875; - ddgiDiffuse *= ddgiSurfaceFactor; - float ddgiDiffuseLuma = dot(ddgiDiffuse, float3(0.2126f, 0.7152f, 0.0722f)); - float sceneRefLuma = - dot(ambientBase + lighting * 0.35f, float3(0.2126f, 0.7152f, 0.0722f)); - float ddgiLumaCap = - sceneRefLuma * 0.85000002384185791015625 + 0.07999999821186065673828125; - if (ddgiDiffuseLuma > ddgiLumaCap) { - ddgiDiffuse *= (ddgiLumaCap / ddgiDiffuseLuma); - } - ddgiDiffuse = max(ddgiDiffuse, float3(0.0f)); + float3 ddgiDiffuse = ddgiIrradiance * albedo * INV_PI * + (1.0f - metallic) * occlusion * ddgiGain; ambient += ddgiDiffuse; float3 ddgiSpecular = float3(0.0f); - if (ps.atlasParams.w > 0.0f && roughness < 0.7f) { - float3 reflectionDir = reflect(-V, N); - float3 ddgiReflection = sampleDDGIIrradiance( - irradianceMap, ps, ddgiSamplePos, reflectionDir); - if (!all(isfinite(ddgiReflection))) { - ddgiReflection = float3(0.0f); - } - ddgiReflection = max(ddgiReflection, float3(0.0f)); - float3 Fddgi = fresnelSchlick(fast::max(dot(N, V), 0.0), F0); - float specGain = mix(0.03f, 0.004f, roughness); - ddgiReflection *= (0.5f + ddgiSurfaceFactor * 0.5f); - ddgiSpecular = ddgiReflection * Fddgi * specGain * INV_PI * ddgiGain; - } float3 iblContribution = float3(0.0); if (_526.useIBL != 0u) { diff --git a/include/photon/illuminate.h b/include/photon/illuminate.h index b7e4f6b6..169af156 100644 --- a/include/photon/illuminate.h +++ b/include/photon/illuminate.h @@ -156,6 +156,8 @@ class GlobalIllumination { std::shared_ptr irradianceMap; /** @brief Previous irradiance atlas for temporal blending. */ std::shared_ptr irradianceMapPrev; + std::shared_ptr distanceMap; + std::shared_ptr distanceMapPrev; /** @brief Compute program that writes probe irradiance tiles to the atlas. */ std::shared_ptr giWriteShader; @@ -171,6 +173,12 @@ class GlobalIllumination { /** @brief Byte capacity currently allocated for probeRadianceBuffer. */ int probeRadianceCapacity = 0; + std::shared_ptr triangleBuffer; + std::shared_ptr materialBuffer; + std::shared_ptr sceneBLAS; + std::shared_ptr sceneTLAS; + bool accelerationStructureDirty = false; + /** @brief Active probe-space definition used for DDGI dispatch. */ std::shared_ptr probeSpace; diff --git a/photon/gi.cpp b/photon/gi.cpp index 3c7d88fd..10b5d35f 100644 --- a/photon/gi.cpp +++ b/photon/gi.cpp @@ -29,7 +29,44 @@ #ifdef METAL namespace { constexpr int kDdgiMaterialTextureUnitStart = 10; -constexpr int kDdgiMaxMaterialTextures = 24; +constexpr int kDdgiMaxMaterialTextures = 48; +constexpr int kDdgiSkyboxTextureUnit = 60; +constexpr int kDdgiPreviousIrradianceTextureUnit = 61; + +std::shared_ptr createDdgiFallbackSkyboxTexture() { + constexpr unsigned char value[4] = {0, 0, 0, 255}; + const unsigned char *faces[6] = {value, value, value, value, value, value}; + auto texture = opal::Texture::create( + opal::TextureType::TextureCubeMap, opal::TextureFormat::Rgba8, 1, 1, + opal::TextureDataFormat::Rgba, nullptr, 1); + texture->setFilterMode(opal::TextureFilterMode::Linear, + opal::TextureFilterMode::Linear); + texture->setWrapMode(opal::TextureAxis::S, + opal::TextureWrapMode::ClampToEdge); + texture->setWrapMode(opal::TextureAxis::T, + opal::TextureWrapMode::ClampToEdge); + texture->setWrapMode(opal::TextureAxis::R, + opal::TextureWrapMode::ClampToEdge); + for (int face = 0; face < 6; ++face) { + texture->updateFace(face, faces[face], 1, 1, + opal::TextureDataFormat::Rgba); + } + return texture; +} + +void clearDdgiTexture(const std::shared_ptr &texture) { + if (texture == nullptr || texture->texture == nullptr || + texture->creationData.width <= 0 || texture->creationData.height <= 0) { + return; + } + std::vector zeros( + static_cast(texture->creationData.width) * + static_cast(texture->creationData.height) * 4, + 0.0f); + texture->texture->updateData(zeros.data(), texture->creationData.width, + texture->creationData.height, + opal::TextureDataFormat::Rgba); +} int registerMaterialTextureSlot( const Texture &texture, @@ -251,6 +288,19 @@ void photon::GlobalIllumination::init() { Texture::create(512, 512, opal::TextureFormat::Rgba16F, opal::TextureDataFormat::Rgba, TextureType::Color)); + distanceMap = std::make_shared( + Texture::create(512, 512, opal::TextureFormat::Rgba16F, + opal::TextureDataFormat::Rgba, TextureType::Color)); + + distanceMapPrev = std::make_shared( + Texture::create(512, 512, opal::TextureFormat::Rgba16F, + opal::TextureDataFormat::Rgba, TextureType::Color)); + + clearDdgiTexture(irradianceMap); + clearDdgiTexture(irradianceMapPrev); + clearDdgiTexture(distanceMap); + clearDdgiTexture(distanceMapPrev); + giPipeline = opal::Pipeline::create(); giPipeline->setShaderProgram(giWriteShader->shader); giPipeline->setComputeThreadgroupSize(8, 8, 1); @@ -299,7 +349,9 @@ void photon::GlobalIllumination::updateProbeLayout() { probeSpace->probeResolution, probeSpace->textureBorderSize); if (hasCachedLayoutSignature && cachedLayoutSignature == layoutSignature && probeRadianceBuffer != nullptr && irradianceMap != nullptr && - irradianceMapPrev != nullptr) { + irradianceMapPrev != nullptr && distanceMap != nullptr && + distanceMapPrev != nullptr && triangleBuffer != nullptr && + materialBuffer != nullptr && sceneBLAS != nullptr) { return; } cachedLayoutSignature = layoutSignature; @@ -311,6 +363,14 @@ void photon::GlobalIllumination::updateProbeLayout() { materials.reserve(ddgiObjects.size()); std::unordered_map textureSlots; + for (auto *object : ddgiObjects) { + if (object == nullptr || !object->canUseDeferredRendering()) { + continue; + } + findTextureSlotForType(object->textures, TextureType::Color, + materialTextures, textureSlots); + } + for (auto *object : ddgiObjects) { if (object == nullptr || !object->canUseDeferredRendering()) { continue; @@ -323,11 +383,7 @@ void photon::GlobalIllumination::updateProbeLayout() { const auto &indices = object->indices; const bool useIndexBuffer = indices.size() >= 3; - glm::mat4 model(1.0f); - model = glm::translate(model, object->getPosition().toGlm()); - model *= - glm::mat4_cast(glm::normalize(object->getRotation().toGlmQuat())); - model = glm::scale(model, object->getScale().toGlm()); + glm::mat4 model = object->model; glm::mat3 linearMatrix = glm::mat3(model); glm::mat3 normalMatrix = glm::transpose(glm::inverse(glm::mat3(model))); @@ -458,6 +514,48 @@ void photon::GlobalIllumination::updateProbeLayout() { } } + if (hasGeometry) { + glm::vec3 rawExtent = glm::max(boundsMax - boundsMin, glm::vec3(0.0f)); + spacing = std::max(spacing, std::max(rawExtent.x, std::max(rawExtent.y, + rawExtent.z)) / + 15.0f); + } + + std::vector accelerationVertices; + std::vector accelerationIndices; + accelerationVertices.reserve(triangles.size() * 3); + accelerationIndices.reserve(triangles.size() * 3); + for (const auto &triangle : triangles) { + const glm::vec4 positions[3] = {triangle.v0, triangle.v1, triangle.v2}; + for (const auto &position : positions) { + opal::PrimitiveVertex vertex{}; + vertex.position[0] = position.x; + vertex.position[1] = position.y; + vertex.position[2] = position.z; + accelerationIndices.push_back( + static_cast(accelerationVertices.size())); + accelerationVertices.push_back(vertex); + } + } + if (!accelerationVertices.empty()) { + sceneBLAS = opal::PrimitiveAccelerationStructure::create( + accelerationVertices, accelerationIndices); + sceneTLAS.reset(); + accelerationStructureDirty = true; + triangleBuffer = opal::Buffer::create( + opal::BufferUsage::ShaderRead, + triangles.size() * sizeof(DDGITriangle), triangles.data()); + materialBuffer = opal::Buffer::create( + opal::BufferUsage::ShaderRead, + materials.size() * sizeof(DDGIMaterial), materials.data()); + } else { + sceneBLAS.reset(); + sceneTLAS.reset(); + triangleBuffer.reset(); + materialBuffer.reset(); + accelerationStructureDirty = false; + } + float layoutPad = spacing * 0.25f; Position3d minWs = hasGeometry ? Position3d(boundsMin.x - layoutPad, boundsMin.y - layoutPad, @@ -498,7 +596,10 @@ void photon::GlobalIllumination::updateProbeLayout() { static_cast(std::ceil(std::sqrt((float)totalProbeCount))), 1, 64); probeSpace->originWorldSpace = minWs; - probeSpace->spacing = Position3d(spacing, spacing, spacing); + probeSpace->spacing = Position3d( + Nx > 1 ? extent.x / static_cast(Nx - 1) : spacing, + Ny > 1 ? extent.y / static_cast(Ny - 1) : spacing, + Nz > 1 ? extent.z / static_cast(Nz - 1) : spacing); probeSpace->probeCount = Vector3((float)Nx, (float)Ny, (float)Nz); probeSpace->probesPerRow = probesPerRow; @@ -522,7 +623,8 @@ void photon::GlobalIllumination::updateProbeLayout() { const int atlasW = std::max(1, probeSpace->atlasWidth()); const int atlasH = std::max(1, probeSpace->atlasHeight()); - bool needCreate = !irradianceMap || !irradianceMapPrev; + bool needCreate = !irradianceMap || !irradianceMapPrev || !distanceMap || + !distanceMapPrev; bool sizeChanged = !needCreate && (irradianceMap->creationData.width != atlasW || irradianceMap->creationData.height != atlasH); @@ -535,6 +637,16 @@ void photon::GlobalIllumination::updateProbeLayout() { irradianceMapPrev = std::make_shared( Texture::create(atlasW, atlasH, opal::TextureFormat::Rgba16F, opal::TextureDataFormat::Rgba, TextureType::Color)); + distanceMap = std::make_shared( + Texture::create(atlasW, atlasH, opal::TextureFormat::Rgba16F, + opal::TextureDataFormat::Rgba, TextureType::Color)); + distanceMapPrev = std::make_shared( + Texture::create(atlasW, atlasH, opal::TextureFormat::Rgba16F, + opal::TextureDataFormat::Rgba, TextureType::Color)); + clearDdgiTexture(irradianceMap); + clearDdgiTexture(irradianceMapPrev); + clearDdgiTexture(distanceMap); + clearDdgiTexture(distanceMapPrev); } int effectiveRaysPerProbe = std::max(1, raysPerProbe); @@ -560,19 +672,36 @@ void photon::GlobalIllumination::render( probeRadianceBuffer == nullptr || irradianceMap == nullptr || irradianceMapPrev == nullptr || irradianceMap->texture == nullptr || irradianceMapPrev->texture == nullptr || + distanceMap == nullptr || distanceMapPrev == nullptr || + distanceMap->texture == nullptr || distanceMapPrev->texture == nullptr || + triangleBuffer == nullptr || materialBuffer == nullptr || + sceneBLAS == nullptr || copySrcFramebuffer == nullptr || copyDstFramebuffer == nullptr) { return; } + if (accelerationStructureDirty) { + commandBuffer->buildPrimitiveAccelerationStructure(sceneBLAS); + opal::AccelerationStructureInstance instance{}; + instance.blas = sceneBLAS; + instance.transform = glm::mat4(1.0f); + instance.instanceId = 0; + instance.mask = 0xFF; + instance.cullDisable = true; + sceneTLAS = opal::InstanceAccelerationStructure::create({instance}); + commandBuffer->buildInstanceAccelerationStructure(sceneTLAS); + accelerationStructureDirty = false; + } + if (sceneTLAS == nullptr || !sceneTLAS->isBuilt) { + return; + } + const uint totalProbes = static_cast(std::max(1, this->probeSpace->totalProbes())); const uint requestedRays = static_cast(std::max(1, this->raysPerProbe)); const uint effectiveRays = std::max(1u, requestedRays); uint updateStride = static_cast(std::max(1, this->probeUpdateStride)); - if (frameIndex < static_cast(updateStride) + 2) { - updateStride = 1u; - } uint updateOffset = (updateStride > 1u) ? static_cast(std::max(0, frameIndex)) % updateStride @@ -597,6 +726,11 @@ void photon::GlobalIllumination::render( auto copy = opal::ResolveAction::createForColorAttachment( copySrcFramebuffer, copyDstFramebuffer, 0); commandBuffer->performResolve(copy); + copySrcFramebuffer->attachTexture(distanceMap->texture, 0); + copyDstFramebuffer->attachTexture(distanceMapPrev->texture, 0); + copy = opal::ResolveAction::createForColorAttachment( + copySrcFramebuffer, copyDstFramebuffer, 0); + commandBuffer->performResolve(copy); // Perform Ray Tracing giRaytracingPipeline->bindShaderReadWriteBuffer("probeRadianceOut", @@ -754,23 +888,8 @@ void photon::GlobalIllumination::render( areaLights.empty() ? sizeof(GPUAreaLight) : areaLights.size() * sizeof(GPUAreaLight)); - DDGITriangle fallbackTriangle{}; - DDGIMaterial fallbackMaterial{}; - const void *triangleData = - triangles.empty() ? static_cast(&fallbackTriangle) - : static_cast(triangles.data()); - const size_t triangleSize = triangles.empty() - ? sizeof(DDGITriangle) - : triangles.size() * sizeof(DDGITriangle); - const void *materialData = - materials.empty() ? static_cast(&fallbackMaterial) - : static_cast(materials.data()); - const size_t materialSize = materials.empty() - ? sizeof(DDGIMaterial) - : materials.size() * sizeof(DDGIMaterial); - giRaytracingPipeline->bindBufferData("tris", triangleData, triangleSize); - giRaytracingPipeline->bindBufferData("materials", materialData, - materialSize); + giRaytracingPipeline->bindBuffer("tris", triangleBuffer, 1); + giRaytracingPipeline->bindBuffer("materials", materialBuffer, 2); struct GPUSceneCounts { uint32_t triCount; @@ -805,6 +924,31 @@ void photon::GlobalIllumination::render( kDdgiMaterialTextureUnitStart + i); } + static std::shared_ptr fallbackSkybox = nullptr; + if (fallbackSkybox == nullptr) { + fallbackSkybox = createDdgiFallbackSkyboxTexture(); + } + std::shared_ptr skyboxTexture = fallbackSkybox; + int useSkybox = 0; + glm::vec3 skyColor(0.12f, 0.14f, 0.18f); + if (scene != nullptr) { + auto skybox = scene->getSkybox(); + if (skybox != nullptr && skybox->cubemap.texture != nullptr) { + skyboxTexture = skybox->cubemap.texture; + useSkybox = 1; + } else if (scene->atmosphere.isEnabled()) { + Color atmosphereColor = scene->atmosphere.getLightColor(); + skyColor = glm::vec3(atmosphereColor.r, atmosphereColor.g, + atmosphereColor.b) * + std::max(scene->atmosphere.getLightIntensity(), 0.0f); + } + } + giRaytracingPipeline->bindTexture("skybox", skyboxTexture, + kDdgiSkyboxTextureUnit); + giRaytracingPipeline->bindTexture("previousIrradiance", + irradianceMapPrev->texture, + kDdgiPreviousIrradianceTextureUnit); + giRaytracingPipeline->setUniform3f( "ps.origin", probeSpace->originWorldSpace.x, probeSpace->originWorldSpace.y, probeSpace->originWorldSpace.z); @@ -837,8 +981,12 @@ void photon::GlobalIllumination::render( static_cast(updateStride)); giRaytracingPipeline->setUniform1i("rt.probeUpdateCount", static_cast(activeProbeCount)); + giRaytracingPipeline->setUniform3f("rt.skyColor", skyColor.x, skyColor.y, + skyColor.z); + giRaytracingPipeline->setUniform1i("rt.useSkybox", useSkybox); commandBuffer->bindPipeline(giRaytracingPipeline); + commandBuffer->bindInstanceAccelerationStructure(sceneTLAS, 10); commandBuffer->dispatch(totalRays, 1, 1); commandBuffer->computeBarrier(); @@ -846,6 +994,8 @@ void photon::GlobalIllumination::render( // Write to irradiance texture giPipeline->bindTexture("outTexture", irradianceMap->texture, 0); giPipeline->bindTexture("prevTexture", irradianceMapPrev->texture, 1); + giPipeline->bindTexture("outDistance", distanceMap->texture, 2); + giPipeline->bindTexture("prevDistance", distanceMapPrev->texture, 3); giPipeline->bindBuffer("probeRadiance", this->probeRadianceBuffer); diff --git a/shaders/metal/deferred/light.frag.metal b/shaders/metal/deferred/light.frag.metal index 2c56bdce..a10f9aed 100644 --- a/shaders/metal/deferred/light.frag.metal +++ b/shaders/metal/deferred/light.frag.metal @@ -919,6 +919,7 @@ sampleProbeDirectionalRadiance(texture2d ddgiTexture, } static inline float3 sampleDDGIIrradiance(texture2d ddgiTexture, + texture2d ddgiDistance, constant ProbeSpace &ps, float3 posWS, float3 normalWS) { uint3 counts = uint3((uint)ps.probeCount.x, (uint)ps.probeCount.y, @@ -955,13 +956,27 @@ static inline float3 sampleDDGIIrradiance(texture2d ddgiTexture, uint pIndex = probeIndexFromCoord(nearest, counts); float4 nearestSample = sampleProbeDirectionalRadiance( ddgiTexture, ps, pIndex, atlasW, atlasH, safeNormal); + float3 nearestProbePos = ps.origin + float3(nearest) * safeSpacing; + float3 nearestDirection = safeNormalizeDDGI( + posWS - nearestProbePos, safeNormal); + float4 nearestMoments = sampleProbeDirectionalRadiance( + ddgiDistance, ps, pIndex, atlasW, atlasH, nearestDirection); + float nearestDistance = length(posWS - nearestProbePos); + float nearestVariance = max(nearestMoments.y - nearestMoments.x * + nearestMoments.x, + 0.0001f); + float nearestDelta = max(nearestDistance - nearestMoments.x, 0.0f); + float nearestVisibility = nearestDelta <= 0.0f + ? 1.0f + : nearestVariance / + (nearestVariance + nearestDelta * + nearestDelta); float nearestValidity = isfinite(nearestSample.w) ? clamp(nearestSample.w, 0.0f, 1.0f) : 0.0f; - nearestValidity = mix(0.05f, 1.0f, nearestValidity); float3 nearestIrr = all(isfinite(nearestSample.xyz)) ? nearestSample.xyz : float3(0.0f); - nearestIrr *= nearestValidity; + nearestIrr *= nearestValidity * nearestVisibility; return max(nearestIrr, float3(0.0f)); } @@ -1011,20 +1026,33 @@ static inline float3 sampleDDGIIrradiance(texture2d ddgiTexture, float4 irrSample = sampleProbeDirectionalRadiance( ddgiTexture, ps, pIndex, atlasW, atlasH, safeNormal); + float3 probeToSurface = -surfaceToProbe; + float4 distanceSample = sampleProbeDirectionalRadiance( + ddgiDistance, ps, pIndex, atlasW, atlasH, + safeNormalizeDDGI(probeToSurface, safeNormal)); + float variance = max(distanceSample.y - + distanceSample.x * distanceSample.x, + 0.0001f); + float delta = max(sDist - distanceSample.x, 0.0f); + float visibility = delta <= 0.0f + ? 1.0f + : variance / (variance + delta * delta); + visibility = visibility * visibility * visibility; float probeValidity = isfinite(irrSample.w) ? clamp(irrSample.w, 0.0f, 1.0f) : 0.0f; - float validityW = mix(0.05f, 1.0f, probeValidity); + float validityW = probeValidity; float3 irr = irrSample.xyz; if (!all(isfinite(irr))) { irr = float3(0.0f); validityW = 0.0f; } irr = max(irr, float3(0.0f)); - float weightedW = w * validityW; + float weightedW = w * validityW * max(visibility, 0.001f); result += irr * weightedW; weightSum += weightedW; - float noBackfaceW = trilinearW * distanceW * validityW; + float noBackfaceW = trilinearW * distanceW * validityW * + max(visibility, 0.001f); resultNoBackface += irr * noBackfaceW; weightSumNoBackface += noBackfaceW; } @@ -1047,7 +1075,6 @@ static inline float3 sampleDDGIIrradiance(texture2d ddgiTexture, ddgiTexture, ps, nearestIndex, atlasW, atlasH, safeNormal); float fallbackValidity = isfinite(fallbackSample.w) ? clamp(fallbackSample.w, 0.0f, 1.0f) : 0.0f; - fallbackValidity = mix(0.05f, 1.0f, fallbackValidity); float3 fallback = all(isfinite(fallbackSample.xyz)) ? fallbackSample.xyz : float3(0.0f); fallback *= fallbackValidity; @@ -1085,6 +1112,7 @@ fragment main0_out main0( texture2d gMaterial [[texture(14)]], texture2d ssao [[texture(15)]], texture2d irradianceMap [[texture(16)]], + texture2d ddgiDistanceMap [[texture(17)]], sampler texture1Smplr [[sampler(0)]], sampler texture2Smplr [[sampler(1)]], sampler texture3Smplr [[sampler(2)]], sampler texture4Smplr [[sampler(3)]], sampler texture5Smplr [[sampler(4)]], sampler cubeMap1Smplr [[sampler(5)]], @@ -1408,13 +1436,15 @@ fragment main0_out main0( float3 ambientBase = ((ambientLight.color.xyz * ambientLight.intensity) * albedo) * occlusion; - float3 ambient = ambientBase; + bool ddgiEnabled = ps.atlasParams.w > 0.0f; + float3 ambient = ddgiEnabled ? ambientBase * 0.05f : ambientBase; float ddgiSampleBias = max(max(ps.spacing.x, max(ps.spacing.y, ps.spacing.z)) * 0.05f, 0.002f); float3 ddgiSamplePos = FragPos + ddgiNormal * ddgiSampleBias; float3 ddgiIrradiance = - sampleDDGIIrradiance(irradianceMap, ps, ddgiSamplePos, ddgiNormal); + sampleDDGIIrradiance(irradianceMap, ddgiDistanceMap, ps, ddgiSamplePos, + ddgiNormal); if (!all(isfinite(ddgiIrradiance))) { ddgiIrradiance = float3(0.0f); } @@ -1427,45 +1457,11 @@ fragment main0_out main0( } const float INV_PI = 0.31830988618379067153776752674503; - float ddgiLuma = dot(ddgiIrradiance, float3(0.2126f, 0.7152f, 0.0722f)); - float3 ddgiChroma = ddgiIrradiance - float3(ddgiLuma); - float3 boostedIrradiance = - max(float3(ddgiLuma * 0.1500000059604644775390625) + - ddgiChroma * 1.35000002384185791015625, - float3(0.0f)); - float3 bleedAlbedo = albedo; - float3 ddgiDiffuse = boostedIrradiance * bleedAlbedo * INV_PI * - (1.0f - metallic) * ddgiGain * - 0.85000002384185791015625; - float sideFactor = clamp(1.0f - abs(N.y), 0.0f, 1.0f); - float ddgiSurfaceFactor = - 0.550000011920928955078125 + sideFactor * 0.449999988079071044921875; - ddgiDiffuse *= ddgiSurfaceFactor; - float ddgiDiffuseLuma = dot(ddgiDiffuse, float3(0.2126f, 0.7152f, 0.0722f)); - float sceneRefLuma = - dot(ambientBase + lighting * 0.35f, float3(0.2126f, 0.7152f, 0.0722f)); - float ddgiLumaCap = - sceneRefLuma * 0.85000002384185791015625 + 0.07999999821186065673828125; - if (ddgiDiffuseLuma > ddgiLumaCap) { - ddgiDiffuse *= (ddgiLumaCap / ddgiDiffuseLuma); - } - ddgiDiffuse = max(ddgiDiffuse, float3(0.0f)); + float3 ddgiDiffuse = ddgiIrradiance * albedo * INV_PI * + (1.0f - metallic) * occlusion * ddgiGain; ambient += ddgiDiffuse; float3 ddgiSpecular = float3(0.0f); - if (ps.atlasParams.w > 0.0f && roughness < 0.7f) { - float3 reflectionDir = reflect(-V, N); - float3 ddgiReflection = sampleDDGIIrradiance( - irradianceMap, ps, ddgiSamplePos, reflectionDir); - if (!all(isfinite(ddgiReflection))) { - ddgiReflection = float3(0.0f); - } - ddgiReflection = max(ddgiReflection, float3(0.0f)); - float3 Fddgi = fresnelSchlick(fast::max(dot(N, V), 0.0), F0); - float specGain = mix(0.03f, 0.004f, roughness); - ddgiReflection *= (0.5f + ddgiSurfaceFactor * 0.5f); - ddgiSpecular = ddgiReflection * Fddgi * specGain * INV_PI * ddgiGain; - } float3 iblContribution = float3(0.0); if (_526.useIBL != 0u) { diff --git a/shaders/metal/gi/ddgi.metal b/shaders/metal/gi/ddgi.metal index 20067d48..3d4b4c74 100644 --- a/shaders/metal/gi/ddgi.metal +++ b/shaders/metal/gi/ddgi.metal @@ -1,5 +1,7 @@ #include +#include using namespace metal; +using namespace raytracing; struct ProbeSpace { float3 origin; @@ -26,6 +28,8 @@ struct RaytracingSettings { uint probeUpdateOffset; uint probeUpdateStride; uint probeUpdateCount; + float3 skyColor; + uint useSkybox; }; struct Material { @@ -179,7 +183,8 @@ static inline bool rayTriangleMT(float3 ro, float3 rd, float3 v0, float3 v1, } static inline Hit traceScene(float3 ro, float3 rd, device const Triangle *tris, - uint triCount) { + instance_acceleration_structure sceneAS, + float maxDistance) { Hit best; best.t = INFINITY; best.hit = 0u; @@ -191,35 +196,36 @@ static inline Hit traceScene(float3 ro, float3 rd, device const Triangle *tris, best.triIndex = 0u; best._pad0 = 0u; - for (uint i = 0; i < triCount; i++) { - float t, u, v; - if (rayTriangleMT(ro, rd, tris[i].v0.xyz, tris[i].v1.xyz, - tris[i].v2.xyz, t, u, v)) { - if (t < best.t) { - best.t = t; - float w = 1.0f - u - v; - float3 n = normalize(tris[i].n0.xyz * w + tris[i].n1.xyz * u + - tris[i].n2.xyz * v); - best.n = n; - best.uv = - tris[i].uv0.xy * w + tris[i].uv1.xy * u + tris[i].uv2.xy * v; - float3 tRaw = tris[i].t0.xyz * w + tris[i].t1.xyz * u + - tris[i].t2.xyz * v; - float tLen2 = dot(tRaw, tRaw); - best.tangent = (tLen2 > 1e-10f) - ? tRaw * rsqrt(tLen2) - : float3(1.0f, 0.0f, 0.0f); - float3 bRaw = tris[i].b0.xyz * w + tris[i].b1.xyz * u + - tris[i].b2.xyz * v; - float bLen2 = dot(bRaw, bRaw); - best.bitangent = (bLen2 > 1e-10f) - ? bRaw * rsqrt(bLen2) - : float3(0.0f, 0.0f, 1.0f); - best.materialID = tris[i].materialID; - best.triIndex = i; - best.hit = 1u; - } - } + intersector isect; + isect.assume_geometry_type(geometry_type::triangle); + isect.set_triangle_cull_mode(triangle_cull_mode::none); + ray query; + query.origin = ro; + query.direction = rd; + query.min_distance = 0.0001f; + query.max_distance = max(maxDistance, query.min_distance); + auto intersection = isect.intersect(query, sceneAS, 0xFF); + if (intersection.type != intersection_type::none) { + uint i = intersection.primitive_id; + float2 bary = intersection.triangle_barycentric_coord; + float u = bary.x; + float v = bary.y; + float w = 1.0f - u - v; + best.t = intersection.distance; + best.n = normalize(tris[i].n0.xyz * w + tris[i].n1.xyz * u + + tris[i].n2.xyz * v); + best.uv = tris[i].uv0.xy * w + tris[i].uv1.xy * u + tris[i].uv2.xy * v; + float3 tRaw = tris[i].t0.xyz * w + tris[i].t1.xyz * u + tris[i].t2.xyz * v; + float tLen2 = dot(tRaw, tRaw); + best.tangent = (tLen2 > 1e-10f) ? tRaw * rsqrt(tLen2) + : float3(1.0f, 0.0f, 0.0f); + float3 bRaw = tris[i].b0.xyz * w + tris[i].b1.xyz * u + tris[i].b2.xyz * v; + float bLen2 = dot(bRaw, bRaw); + best.bitangent = (bLen2 > 1e-10f) ? bRaw * rsqrt(bLen2) + : float3(0.0f, 0.0f, 1.0f); + best.materialID = tris[i].materialID; + best.triIndex = i; + best.hit = 1u; } return best; @@ -272,7 +278,19 @@ static inline float4 sampleMaterialTexture( texture2d materialTexture17, texture2d materialTexture18, texture2d materialTexture19, texture2d materialTexture20, texture2d materialTexture21, texture2d materialTexture22, - texture2d materialTexture23) { + texture2d materialTexture23, texture2d materialTexture24, + texture2d materialTexture25, texture2d materialTexture26, + texture2d materialTexture27, texture2d materialTexture28, + texture2d materialTexture29, texture2d materialTexture30, + texture2d materialTexture31, texture2d materialTexture32, + texture2d materialTexture33, texture2d materialTexture34, + texture2d materialTexture35, texture2d materialTexture36, + texture2d materialTexture37, texture2d materialTexture38, + texture2d materialTexture39, texture2d materialTexture40, + texture2d materialTexture41, texture2d materialTexture42, + texture2d materialTexture43, texture2d materialTexture44, + texture2d materialTexture45, texture2d materialTexture46, + texture2d materialTexture47) { switch (textureIndex) { case 0: return materialTexture0.sample(materialTexSampler, uv); @@ -322,6 +340,54 @@ static inline float4 sampleMaterialTexture( return materialTexture22.sample(materialTexSampler, uv); case 23: return materialTexture23.sample(materialTexSampler, uv); + case 24: + return materialTexture24.sample(materialTexSampler, uv); + case 25: + return materialTexture25.sample(materialTexSampler, uv); + case 26: + return materialTexture26.sample(materialTexSampler, uv); + case 27: + return materialTexture27.sample(materialTexSampler, uv); + case 28: + return materialTexture28.sample(materialTexSampler, uv); + case 29: + return materialTexture29.sample(materialTexSampler, uv); + case 30: + return materialTexture30.sample(materialTexSampler, uv); + case 31: + return materialTexture31.sample(materialTexSampler, uv); + case 32: + return materialTexture32.sample(materialTexSampler, uv); + case 33: + return materialTexture33.sample(materialTexSampler, uv); + case 34: + return materialTexture34.sample(materialTexSampler, uv); + case 35: + return materialTexture35.sample(materialTexSampler, uv); + case 36: + return materialTexture36.sample(materialTexSampler, uv); + case 37: + return materialTexture37.sample(materialTexSampler, uv); + case 38: + return materialTexture38.sample(materialTexSampler, uv); + case 39: + return materialTexture39.sample(materialTexSampler, uv); + case 40: + return materialTexture40.sample(materialTexSampler, uv); + case 41: + return materialTexture41.sample(materialTexSampler, uv); + case 42: + return materialTexture42.sample(materialTexSampler, uv); + case 43: + return materialTexture43.sample(materialTexSampler, uv); + case 44: + return materialTexture44.sample(materialTexSampler, uv); + case 45: + return materialTexture45.sample(materialTexSampler, uv); + case 46: + return materialTexture46.sample(materialTexSampler, uv); + case 47: + return materialTexture47.sample(materialTexSampler, uv); default: break; } @@ -342,7 +408,19 @@ static inline void resolveMaterialParameters( texture2d materialTexture17, texture2d materialTexture18, texture2d materialTexture19, texture2d materialTexture20, texture2d materialTexture21, texture2d materialTexture22, - texture2d materialTexture23, thread float3 &albedo, + texture2d materialTexture23, texture2d materialTexture24, + texture2d materialTexture25, texture2d materialTexture26, + texture2d materialTexture27, texture2d materialTexture28, + texture2d materialTexture29, texture2d materialTexture30, + texture2d materialTexture31, texture2d materialTexture32, + texture2d materialTexture33, texture2d materialTexture34, + texture2d materialTexture35, texture2d materialTexture36, + texture2d materialTexture37, texture2d materialTexture38, + texture2d materialTexture39, texture2d materialTexture40, + texture2d materialTexture41, texture2d materialTexture42, + texture2d materialTexture43, texture2d materialTexture44, + texture2d materialTexture45, texture2d materialTexture46, + texture2d materialTexture47, thread float3 &albedo, thread float &metallic, thread float &roughness, thread float &ao, thread float3 &emissive, thread int &normalTextureIndex, thread float &normalStrength) { @@ -380,7 +458,19 @@ static inline void resolveMaterialParameters( materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, - materialTexture22, materialTexture23) + materialTexture22, materialTexture23, + materialTexture24, materialTexture25, + materialTexture26, materialTexture27, + materialTexture28, materialTexture29, + materialTexture30, materialTexture31, + materialTexture32, materialTexture33, + materialTexture34, materialTexture35, + materialTexture36, materialTexture37, + materialTexture38, materialTexture39, + materialTexture40, materialTexture41, + materialTexture42, materialTexture43, + materialTexture44, materialTexture45, + materialTexture46, materialTexture47) .xyz, float3(0.0f), float3(1.0f)); } @@ -395,7 +485,15 @@ static inline void resolveMaterialParameters( materialTexture14, materialTexture15, materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, materialTexture22, - materialTexture23); + materialTexture23, materialTexture24, materialTexture25, + materialTexture26, materialTexture27, materialTexture28, + materialTexture29, materialTexture30, materialTexture31, + materialTexture32, materialTexture33, materialTexture34, + materialTexture35, materialTexture36, materialTexture37, + materialTexture38, materialTexture39, materialTexture40, + materialTexture41, materialTexture42, materialTexture43, + materialTexture44, materialTexture45, materialTexture46, + materialTexture47); float metallicValue = metallicSample.x; if (mat.roughnessTextureIndex == mat.metallicTextureIndex) { metallicValue = metallicSample.z; @@ -413,7 +511,15 @@ static inline void resolveMaterialParameters( materialTexture14, materialTexture15, materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, materialTexture22, - materialTexture23); + materialTexture23, materialTexture24, materialTexture25, + materialTexture26, materialTexture27, materialTexture28, + materialTexture29, materialTexture30, materialTexture31, + materialTexture32, materialTexture33, materialTexture34, + materialTexture35, materialTexture36, materialTexture37, + materialTexture38, materialTexture39, materialTexture40, + materialTexture41, materialTexture42, materialTexture43, + materialTexture44, materialTexture45, materialTexture46, + materialTexture47); float roughnessValue = roughnessSample.x; if (mat.roughnessTextureIndex == mat.metallicTextureIndex) { roughnessValue = roughnessSample.y; @@ -432,7 +538,19 @@ static inline void resolveMaterialParameters( materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, - materialTexture22, materialTexture23) + materialTexture22, materialTexture23, + materialTexture24, materialTexture25, + materialTexture26, materialTexture27, + materialTexture28, materialTexture29, + materialTexture30, materialTexture31, + materialTexture32, materialTexture33, + materialTexture34, materialTexture35, + materialTexture36, materialTexture37, + materialTexture38, materialTexture39, + materialTexture40, materialTexture41, + materialTexture42, materialTexture43, + materialTexture44, materialTexture45, + materialTexture46, materialTexture47) .x, 0.0f, 1.0f); } @@ -457,7 +575,19 @@ static inline float3 resolveNormal( texture2d materialTexture17, texture2d materialTexture18, texture2d materialTexture19, texture2d materialTexture20, texture2d materialTexture21, texture2d materialTexture22, - texture2d materialTexture23) { + texture2d materialTexture23, texture2d materialTexture24, + texture2d materialTexture25, texture2d materialTexture26, + texture2d materialTexture27, texture2d materialTexture28, + texture2d materialTexture29, texture2d materialTexture30, + texture2d materialTexture31, texture2d materialTexture32, + texture2d materialTexture33, texture2d materialTexture34, + texture2d materialTexture35, texture2d materialTexture36, + texture2d materialTexture37, texture2d materialTexture38, + texture2d materialTexture39, texture2d materialTexture40, + texture2d materialTexture41, texture2d materialTexture42, + texture2d materialTexture43, texture2d materialTexture44, + texture2d materialTexture45, texture2d materialTexture46, + texture2d materialTexture47) { float3 N = safeNormalize(n, float3(0.0f, 1.0f, 0.0f)); float3 T = safeNormalize(tangent - N * dot(N, tangent), float3(1.0f, 0.0f, 0.0f)); float3 B = safeNormalize(bitangent - N * dot(N, bitangent), cross(N, T)); @@ -480,7 +610,19 @@ static inline float3 resolveNormal( materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, - materialTexture22, materialTexture23) + materialTexture22, materialTexture23, + materialTexture24, materialTexture25, + materialTexture26, materialTexture27, + materialTexture28, materialTexture29, + materialTexture30, materialTexture31, + materialTexture32, materialTexture33, + materialTexture34, materialTexture35, + materialTexture36, materialTexture37, + materialTexture38, materialTexture39, + materialTexture40, materialTexture41, + materialTexture42, materialTexture43, + materialTexture44, materialTexture45, + materialTexture46, materialTexture47) .xyz; texN = texN * 2.0f - 1.0f; texN.xy *= normalStrength; @@ -490,20 +632,66 @@ static inline float3 resolveNormal( return N; } -static inline float3 sampleSky(float3 d) { +static inline float2 octEncode(float3 n) { + n /= max(fabs(n.x) + fabs(n.y) + fabs(n.z), 1e-6f); + float2 e = n.xy; + if (n.z < 0.0f) { + float2 signNotZero = + float2(e.x >= 0.0f ? 1.0f : -1.0f, e.y >= 0.0f ? 1.0f : -1.0f); + e = (1.0f - fabs(e.yx)) * signNotZero; + } + return e; +} + +constexpr sampler ddgiLinearSampler(coord::normalized, address::clamp_to_edge, + filter::linear); + +static inline float3 samplePreviousIrradiance(texture2d previous, + constant ProbeSpace &ps, + float3 posWS, float3 direction) { + uint3 counts = uint3(ps.probeCount); + if (counts.x == 0u || counts.y == 0u || counts.z == 0u) { + return float3(0.0f); + } + float3 grid = clamp((posWS - ps.origin) / max(ps.spacing, float3(1e-4f)), + float3(0.0f), float3(counts - 1u)); + uint3 coord = uint3(round(grid)); + uint probeIndex = coord.x + counts.x * (coord.y + counts.y * coord.z); + uint border = uint(ps.atlasParams.x); + uint innerRes = uint(ps.atlasParams.y); + uint probesPerRow = uint(ps.atlasParams.z); + uint tileRes = innerRes + 2u * border; + uint2 tile = uint2(probeIndex % probesPerRow, probeIndex / probesPerRow); + float2 inner = octEncode(safeNormalize(direction, float3(0.0f, 1.0f, 0.0f))) * + 0.5f + + 0.5f; + float2 pixel = float2(tile * tileRes + border) + + inner * float(max(innerRes, 1u) - 1u) + 0.5f; + float3 history = previous.sample( + ddgiLinearSampler, + pixel / float2(previous.get_width(), + previous.get_height())) + .xyz; + return all(isfinite(history)) ? max(history, float3(0.0f)) + : float3(0.0f); +} + +static inline float3 sampleSky(float3 d, texturecube skybox, + float3 skyColor, uint useSkybox) { + if (useSkybox != 0u) { + return max(skybox.sample(ddgiLinearSampler, d).xyz, float3(0.0f)); + } float t = clamp(d.y * 0.5f + 0.5f, 0.0f, 1.0f); - return mix(float3(0.02f, 0.023f, 0.028f), - float3(0.12f, 0.14f, 0.18f), - t); + return mix(skyColor * 0.08f, skyColor, t); } static inline float shadowVisibility(float3 ro, float3 rd, float maxT, device const Triangle *tris, - uint triCount) { + instance_acceleration_structure sceneAS) { if (maxT <= 1e-4f) { return 1.0f; } - Hit h = traceScene(ro, rd, tris, triCount); + Hit h = traceScene(ro, rd, tris, sceneAS, maxT); if (!h.hit) { return 1.0f; } @@ -520,7 +708,7 @@ static inline float giNdotL(float3 n, float3 l) { static inline float3 evaluateDirectLights( float3 posWS, float3 normalWS, float bias, float maxDistance, - device const Triangle *tris, uint triCount, + device const Triangle *tris, instance_acceleration_structure sceneAS, device const DirectionalLight *directionalLights, uint directionalLightCount, device const PointLight *pointLights, uint pointLightCount, device const SpotLight *spotLights, uint spotLightCount, @@ -535,8 +723,10 @@ static inline float3 evaluateDirectLights( if (ndl <= 0.0f) continue; + float visibility = shadowVisibility(posWS + n * bias, L, maxDistance, + tris, sceneAS); sum += directionalLights[i].diffuse * - max(0.0f, directionalLights[i].intensity) * ndl; + max(0.0f, directionalLights[i].intensity) * ndl * visibility; } for (uint i = 0; i < pointLightCount; i++) { @@ -556,8 +746,10 @@ static inline float3 evaluateDirectLights( pointLights[i].quadratic * dist * dist, 1e-4f); float fade = 1.0f - smoothstep(radius * 0.9f, radius, dist); + float visibility = shadowVisibility(posWS + n * bias, L, dist - bias, + tris, sceneAS); sum += pointLights[i].diffuse * max(0.0f, pointLights[i].intensity) * - attenuation * fade * ndl; + attenuation * fade * ndl * visibility; } for (uint i = 0; i < spotLightCount; i++) { @@ -585,8 +777,10 @@ static inline float3 evaluateDirectLights( float attenuation = 1.0f / ((1.0f + (dist / range)) + ((dist * dist) / (range * range))); float fade = 1.0f - smoothstep(range * 0.9f, range, dist); + float visibility = shadowVisibility(posWS + n * bias, L, dist - bias, + tris, sceneAS); sum += spotLights[i].diffuse * max(0.0f, spotLights[i].intensity) * cone * - attenuation * fade * ndl; + attenuation * fade * ndl * visibility; } for (uint i = 0; i < areaLightCount; i++) { @@ -622,8 +816,10 @@ static inline float3 evaluateDirectLights( float attenuation = 1.0f / ((1.0f + (dist / range)) + ((dist * dist) / (range * range))); float fade = 1.0f - smoothstep(range * 0.9f, range, dist); + float visibility = shadowVisibility(posWS + n * bias, L, dist - bias, + tris, sceneAS); sum += areaLights[i].diffuse * max(0.0f, areaLights[i].intensity) * - facing * attenuation * fade * ndl; + facing * attenuation * fade * ndl * visibility; } return sum; @@ -670,6 +866,7 @@ kernel void main0(device float4 *probeRadianceOut [[buffer(0)]], device const PointLight *pointLights [[buffer(7)]], device const SpotLight *spotLights [[buffer(8)]], device const AreaLight *areaLights [[buffer(9)]], + instance_acceleration_structure sceneAS [[buffer(10)]], texture2d materialTexture0 [[texture(10)]], texture2d materialTexture1 [[texture(11)]], texture2d materialTexture2 [[texture(12)]], @@ -694,6 +891,32 @@ kernel void main0(device float4 *probeRadianceOut [[buffer(0)]], texture2d materialTexture21 [[texture(31)]], texture2d materialTexture22 [[texture(32)]], texture2d materialTexture23 [[texture(33)]], + texture2d materialTexture24 [[texture(34)]], + texture2d materialTexture25 [[texture(35)]], + texture2d materialTexture26 [[texture(36)]], + texture2d materialTexture27 [[texture(37)]], + texture2d materialTexture28 [[texture(38)]], + texture2d materialTexture29 [[texture(39)]], + texture2d materialTexture30 [[texture(40)]], + texture2d materialTexture31 [[texture(41)]], + texture2d materialTexture32 [[texture(42)]], + texture2d materialTexture33 [[texture(43)]], + texture2d materialTexture34 [[texture(44)]], + texture2d materialTexture35 [[texture(45)]], + texture2d materialTexture36 [[texture(46)]], + texture2d materialTexture37 [[texture(47)]], + texture2d materialTexture38 [[texture(48)]], + texture2d materialTexture39 [[texture(49)]], + texture2d materialTexture40 [[texture(50)]], + texture2d materialTexture41 [[texture(51)]], + texture2d materialTexture42 [[texture(52)]], + texture2d materialTexture43 [[texture(53)]], + texture2d materialTexture44 [[texture(54)]], + texture2d materialTexture45 [[texture(55)]], + texture2d materialTexture46 [[texture(56)]], + texture2d materialTexture47 [[texture(57)]], + texturecube skybox [[texture(60)]], + texture2d previousIrradiance [[texture(61)]], uint tid [[thread_position_in_grid]]) { const float PI = 3.14159265359f; uint totalProbes = (uint)ps.atlasParams.w; @@ -736,7 +959,7 @@ kernel void main0(device float4 *probeRadianceOut [[buffer(0)]], Hit h; float selfHitThreshold = bias * 4.0f; for (uint escapeStep = 0u; escapeStep < 1u; escapeStep++) { - h = traceScene(ro, rayDir, tris, sc.triCount); + h = traceScene(ro, rayDir, tris, sceneAS, maxDistance); if (h.hit == 0u || h.t >= selfHitThreshold) { break; } @@ -749,7 +972,7 @@ kernel void main0(device float4 *probeRadianceOut [[buffer(0)]], float3 radiance = float3(0.0f); if (h.hit == 0u || h.t > maxDistance) { - radiance = sampleSky(rayDir); + radiance = sampleSky(rayDir, skybox, rt.skyColor, rt.useSkybox); } else { float3 hitPos = ro + rayDir * h.t; float3 albedo; @@ -769,7 +992,15 @@ kernel void main0(device float4 *probeRadianceOut [[buffer(0)]], materialTexture13, materialTexture14, materialTexture15, materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, - materialTexture22, materialTexture23, albedo, metallic, roughness, + materialTexture22, materialTexture23, materialTexture24, + materialTexture25, materialTexture26, materialTexture27, + materialTexture28, materialTexture29, materialTexture30, + materialTexture31, materialTexture32, materialTexture33, + materialTexture34, materialTexture35, materialTexture36, + materialTexture37, materialTexture38, materialTexture39, + materialTexture40, materialTexture41, materialTexture42, + materialTexture43, materialTexture44, materialTexture45, + materialTexture46, materialTexture47, albedo, metallic, roughness, ao, emissive, normalTextureIndex, normalStrength); float3 hitNormal = resolveNormal( @@ -782,21 +1013,35 @@ kernel void main0(device float4 *probeRadianceOut [[buffer(0)]], materialTexture14, materialTexture15, materialTexture16, materialTexture17, materialTexture18, materialTexture19, materialTexture20, materialTexture21, materialTexture22, - materialTexture23); + materialTexture23, materialTexture24, materialTexture25, + materialTexture26, materialTexture27, materialTexture28, + materialTexture29, materialTexture30, materialTexture31, + materialTexture32, materialTexture33, materialTexture34, + materialTexture35, materialTexture36, materialTexture37, + materialTexture38, materialTexture39, materialTexture40, + materialTexture41, materialTexture42, materialTexture43, + materialTexture44, materialTexture45, materialTexture46, + materialTexture47); if (dot(hitNormal, -rayDir) < 0.0f) { hitNormal = -hitNormal; } float3 direct = evaluateDirectLights( - hitPos, hitNormal, bias, maxDistance, tris, sc.triCount, + hitPos, hitNormal, bias, maxDistance, tris, sceneAS, directionalLights, sc.directionalLightCount, pointLights, sc.pointLightCount, spotLights, sc.spotLightCount, areaLights, sc.areaLightCount); - float diffuseWeight = - (1.0f - metallic) * mix(0.35f, 1.0f, 1.0f - roughness); - radiance = direct * albedo * diffuseWeight * max(ao, 0.05f) + emissive; + float diffuseWeight = 1.0f - metallic; + float3 diffuseResponse = albedo * diffuseWeight * max(ao, 0.05f) / PI; + float3 previousBounce = + rt.frameIndex >= max(rt.probeUpdateStride, 1u) + ? samplePreviousIrradiance(previousIrradiance, ps, + hitPos + hitNormal * bias, hitNormal) + : float3(0.0f); + float3 indirect = previousBounce * diffuseResponse * 0.35f; + radiance = direct * diffuseResponse + indirect + emissive; radiance = clamp(radiance, float3(0.0f), float3(16.0f)); } diff --git a/shaders/metal/gi/ddgi_write.metal b/shaders/metal/gi/ddgi_write.metal index 11d4d114..a53e07db 100644 --- a/shaders/metal/gi/ddgi_write.metal +++ b/shaders/metal/gi/ddgi_write.metal @@ -64,6 +64,8 @@ static inline float3 sphericalFibonacci(uint index, uint count, uint frameIndex) kernel void main0(texture2d outTexture [[texture(0)]], texture2d prevTexture [[texture(1)]], + texture2d outDistance [[texture(2)]], + texture2d prevDistance [[texture(3)]], device float4 *probeRadiance [[buffer(0)]], constant ProbeSpace &ps [[buffer(1)]], constant RaytracingSettings &rt [[buffer(2)]], @@ -82,6 +84,7 @@ kernel void main0(texture2d outTexture [[texture(0)]], if (tileRes == 0u || probesPerRow == 0u || totalProbes == 0u || innerRes == 0u) { outTexture.write(prevTexture.read(gid), gid); + outDistance.write(prevDistance.read(gid), gid); return; } @@ -91,6 +94,7 @@ kernel void main0(texture2d outTexture [[texture(0)]], if (probeIndex >= totalProbes) { outTexture.write(prevTexture.read(gid), gid); + outDistance.write(prevDistance.read(gid), gid); return; } @@ -103,6 +107,7 @@ kernel void main0(texture2d outTexture [[texture(0)]], (((probeIndex - updateOffset) % updateStride) == 0u)); if (!probeIsActive) { outTexture.write(prevTexture.read(gid), gid); + outDistance.write(prevDistance.read(gid), gid); return; } @@ -122,12 +127,13 @@ kernel void main0(texture2d outTexture [[texture(0)]], float3 sum = float3(0.0f); float weightSum = 0.0f; + float distanceSum = 0.0f; + float distanceSquaredSum = 0.0f; float nearHitCount = 0.0f; - float missCount = 0.0f; float spacingScale = max(max(ps.spacing.x, max(ps.spacing.y, ps.spacing.z)), 1e-4f); float nearHitThreshold = - max(max(rt.normalBias * 1.2f, spacingScale * 0.015f), 0.0008f); + max(max(rt.normalBias * 2.0f, spacingScale * 0.1f), 0.002f); for (uint r = 0; r < raysPerProbe; r += rayStep) { sampledRayCount++; @@ -135,8 +141,6 @@ kernel void main0(texture2d outTexture [[texture(0)]], float hitDistance = raySample.w; if (hitDistance > 0.0f && hitDistance < nearHitThreshold) { nearHitCount += 1.0f; - } else if (hitDistance <= 0.0f) { - missCount += 1.0f; } float3 rayDir = sphericalFibonacci(r, raysPerProbe, rt.frameIndex); @@ -144,19 +148,24 @@ kernel void main0(texture2d outTexture [[texture(0)]], if (w > 1e-6f) { float3 rad = raySample.xyz; if (all(isfinite(rad))) { - float lum = dot(rad, float3(0.2126f, 0.7152f, 0.0722f)); - float compression = 1.0f / (1.0f + lum * 0.25f); - rad *= compression; sum += rad * w; weightSum += w; + float distance = hitDistance > 0.0f + ? min(hitDistance, rt.maxRayDistance) + : rt.maxRayDistance; + distanceSum += distance * w; + distanceSquaredSum += distance * distance * w; } } } float3 irradiance = float3(0.0f); + float2 distanceMoments = float2(rt.maxRayDistance, + rt.maxRayDistance * rt.maxRayDistance); float invRayCount = 1.0f / float(max(sampledRayCount, 1u)); if (weightSum > 1e-6f) { irradiance = sum * (FOUR_PI * invRayCount); + distanceMoments = float2(distanceSum, distanceSquaredSum) / weightSum; } if (!all(isfinite(irradiance))) { @@ -164,22 +173,28 @@ kernel void main0(texture2d outTexture [[texture(0)]], } float4 prev = prevTexture.read(gid); + float4 previousDistance = prevDistance.read(gid); float3 prevValue = all(isfinite(prev.xyz)) ? prev.xyz : float3(0.0f); float prevValidity = isfinite(prev.w) ? clamp(prev.w, 0.0f, 1.0f) : 1.0f; float nearFraction = nearHitCount * invRayCount; - float missFraction = missCount * invRayCount; float nearPenalty = smoothstep(0.82f, 0.995f, nearFraction); - float missPenalty = smoothstep(0.95f, 1.0f, missFraction); - float probeValidity = (1.0f - nearPenalty) * (1.0f - missPenalty); - probeValidity = clamp(probeValidity, 0.005f, 1.0f); + float probeValidity = 1.0f - nearPenalty; + probeValidity = clamp(probeValidity, 0.0f, 1.0f); float h = clamp(rt.hysteresis, 0.0f, 0.995f); - float3 blended = (rt.frameIndex == 0u) ? irradiance : mix(irradiance, prevValue, h); + bool firstProbeUpdate = rt.frameIndex < updateStride; + float3 blended = firstProbeUpdate ? irradiance : mix(irradiance, prevValue, h); + float2 blendedDistance = + firstProbeUpdate + ? distanceMoments + : mix(distanceMoments, previousDistance.xy, h); float blendedValidity = - (rt.frameIndex == 0u) + firstProbeUpdate ? probeValidity : mix(probeValidity, prevValidity, h); outTexture.write(float4(max(blended, float3(0.0f)), blendedValidity), gid); + outDistance.write(float4(max(blendedDistance, float2(0.0f)), + blendedValidity, 0.0f), gid); }