diff --git a/zenovis/xinxinoptix/PTKernel.cu b/zenovis/xinxinoptix/PTKernel.cu index e5ee736af3..2458338ac6 100644 --- a/zenovis/xinxinoptix/PTKernel.cu +++ b/zenovis/xinxinoptix/PTKernel.cu @@ -92,6 +92,32 @@ vec3 PhysicalCamera(vec3 in, mapped = in * exposure; return enableExposure? (enableACES? ACESFilm(mapped):mapped ) : (enableACES? ACESFilm(in) : in); } + +static __inline__ __device__ +ushort3 float3_to_half3(float3 in) +{ + half x = __float2half(in.x); + half y = __float2half(in.y); + half z = __float2half(in.z); + ushort3 v; + v.x = reinterpret_cast(x); + v.y = reinterpret_cast(y); + v.z = reinterpret_cast(z); + return v; +} + +static __inline__ __device__ +float3 half3_to_float3(ushort3 in) +{ + half x = reinterpret_cast(in.x); + half y = reinterpret_cast(in.y); + half z = reinterpret_cast(in.z); + float3 v; + v.x = __half2float(x); + v.y = __half2float(y); + v.z = __half2float(z); + return v; +} extern "C" __global__ void __raygen__rg() { @@ -362,7 +388,7 @@ extern "C" __global__ void __raygen__rg() const float3 accum_color_prev_s = make_float3( params.accum_buffer_S[ image_index ]); const float3 accum_color_prev_t = make_float3( params.accum_buffer_T[ image_index ]); const float3 accum_color_prev_b = make_float3( params.accum_buffer_B[ image_index ]); - const float3 accum_mask_prev = params.frame_buffer_M[ image_index ]; + const float3 accum_mask_prev = half3_to_float3(params.frame_buffer_M[ image_index ]); accum_color = mix( vec3(accum_color_prev), accum_color, a ); accum_color_d = mix( vec3(accum_color_prev_d), accum_color_d, a ); accum_color_s = mix( vec3(accum_color_prev_s), accum_color_s, a ); @@ -399,12 +425,12 @@ extern "C" __global__ void __raygen__rg() float3 out_color_t = t_mapped; float3 out_color_b = accum_color_b; params.frame_buffer[ image_index ] = make_color ( out_color ); - params.frame_buffer_C[ image_index ] = out_color; - params.frame_buffer_D[ image_index ] = out_color_d; - params.frame_buffer_S[ image_index ] = out_color_s; - params.frame_buffer_T[ image_index ] = out_color_t; - params.frame_buffer_B[ image_index ] = accum_color_b; - params.frame_buffer_M[ image_index ] = accum_mask; + params.frame_buffer_C[ image_index ] = float3_to_half3(out_color); + params.frame_buffer_D[ image_index ] = float3_to_half3(out_color_d); + params.frame_buffer_S[ image_index ] = float3_to_half3(out_color_s); + params.frame_buffer_T[ image_index ] = float3_to_half3(out_color_t); + params.frame_buffer_B[ image_index ] = float3_to_half3(accum_color_b); + params.frame_buffer_M[ image_index ] = float3_to_half3(accum_mask); if (params.denoise) { params.albedo_buffer[ image_index ] = tmp_albedo; diff --git a/zenovis/xinxinoptix/optixPathTracer.cpp b/zenovis/xinxinoptix/optixPathTracer.cpp index b33f92b89f..9eeb133087 100644 --- a/zenovis/xinxinoptix/optixPathTracer.cpp +++ b/zenovis/xinxinoptix/optixPathTracer.cpp @@ -253,12 +253,12 @@ ushort2 halfNormal(float4 in) #endif std::optional> output_buffer_o; -std::optional> output_buffer_color; -std::optional> output_buffer_diffuse; -std::optional> output_buffer_specular; -std::optional> output_buffer_transmit; -std::optional> output_buffer_background; -std::optional> output_buffer_mask; +std::optional> output_buffer_color; +std::optional> output_buffer_diffuse; +std::optional> output_buffer_specular; +std::optional> output_buffer_transmit; +std::optional> output_buffer_background; +std::optional> output_buffer_mask; using Vertex = float4; struct PathTracerState @@ -3805,6 +3805,15 @@ void *optixgetimg_extra(std::string name) { } throw std::runtime_error("invalid optixgetimg_extra name: " + name); } + +std::vector optixgetimg_extra(std::string name, int w, int h) { + std::vector data(w * h * 3); + short *ptr = (short*)optixgetimg_extra(name); + for (auto i = 0; i < data.size(); i++) { + data[i] = glm::detail::toFloat32(ptr[i]); + } + return data; +} static void save_exr(float3* ptr, int w, int h, std::string path) { std::vector data(w * h); std::copy_n(ptr, w * h, data.data()); @@ -3879,7 +3888,7 @@ void optixrender(int fbo, int samples, bool denoise, bool simpleRender) { auto exr_path = path.substr(0, path.size() - 4) + ".exr"; if (enable_output_mask) { path = path.substr(0, path.size() - 4); - save_png_data(path + "_mask.png", w, h, (float*)optixgetimg_extra("mask")); + save_png_data(path + "_mask.png", w, h, optixgetimg_extra("mask", w, h).data()); } // AOV if (enable_output_aov) { @@ -3887,12 +3896,12 @@ void optixrender(int fbo, int samples, bool denoise, bool simpleRender) { zeno::create_directories_when_write_file(exr_path); SaveMultiLayerEXR( { - (float*)optixgetimg_extra("color"), - (float*)optixgetimg_extra("diffuse"), - (float*)optixgetimg_extra("specular"), - (float*)optixgetimg_extra("transmit"), - (float*)optixgetimg_extra("background"), - (float*)optixgetimg_extra("mask"), + optixgetimg_extra("color", w, h).data(), + optixgetimg_extra("diffuse", w, h).data(), + optixgetimg_extra("specular", w, h).data(), + optixgetimg_extra("transmit", w, h).data(), + optixgetimg_extra("background", w, h).data(), + optixgetimg_extra("mask", w, h).data(), }, w, h, @@ -3909,18 +3918,20 @@ void optixrender(int fbo, int samples, bool denoise, bool simpleRender) { } else { + zeno::create_directories_when_write_file(exr_path); + save_exr((float3 *)optixgetimg_extra("color", w, h).data(), w, h, exr_path); path = path.substr(0, path.size() - 4); - save_png_color(path + ".aov.diffuse.png", w, h, (float*)optixgetimg_extra("diffuse")); - save_png_color(path + ".aov.specular.png", w, h, (float*)optixgetimg_extra("specular")); - save_png_color(path + ".aov.transmit.png", w, h, (float*)optixgetimg_extra("transmit")); - save_png_data(path + ".aov.background.png", w, h, (float*)optixgetimg_extra("background")); - save_png_data(path + ".aov.mask.png", w, h, (float*)optixgetimg_extra("mask")); + save_png_color(path + ".aov.diffuse.png", w, h, optixgetimg_extra("diffuse", w, h).data()); + save_png_color(path + ".aov.specular.png", w, h, optixgetimg_extra("specular", w, h).data()); + save_png_color(path + ".aov.transmit.png", w, h, optixgetimg_extra("transmit", w, h).data()); + save_png_data(path + ".aov.background.png", w, h, optixgetimg_extra("background", w, h).data()); + save_png_data(path + ".aov.mask.png", w, h, optixgetimg_extra("mask", w, h).data()); } } else { if (enable_output_exr) { zeno::create_directories_when_write_file(exr_path); - save_exr((float3 *)optixgetimg_extra("color"), w, h, exr_path); + save_exr((float3 *)optixgetimg_extra("color", w, h).data(), w, h, exr_path); } else { std::string jpg_native_path = zeno::create_directories_when_write_file(path); diff --git a/zenovis/xinxinoptix/optixPathTracer.h b/zenovis/xinxinoptix/optixPathTracer.h index 7df84d652e..47386835dd 100644 --- a/zenovis/xinxinoptix/optixPathTracer.h +++ b/zenovis/xinxinoptix/optixPathTracer.h @@ -160,12 +160,12 @@ struct Params float4* accum_buffer_T; float4* accum_buffer_B; uchar4* frame_buffer; - float3* frame_buffer_C; - float3* frame_buffer_D; - float3* frame_buffer_S; - float3* frame_buffer_T; - float3* frame_buffer_B; - float3* frame_buffer_M; + ushort3* frame_buffer_C; + ushort3* frame_buffer_D; + ushort3* frame_buffer_S; + ushort3* frame_buffer_T; + ushort3* frame_buffer_B; + ushort3* frame_buffer_M; float3* debug_buffer; float3* albedo_buffer;