Cr8-xyz/blender
1
0
Fork 0
Code Issues Pull Requests Actions 1 Packages Projects Releases Wiki Activity

Cycles: Render Passes

Browse Source 
Currently supported passes:
* Combined, Z, Normal, Object Index, Material Index, Emission, Environment,
  Diffuse/Glossy/Transmission x Direct/Indirect/Color

Not supported yet:
* UV, Vector, Mist

Only enabled for CPU devices at the moment, will do GPU tweaks tommorrow,
also for environment importance sampling.

Documentation:
http://wiki.blender.org/index.php/Doc:2.6/Manual/Render/Cycles/Passes
This commit is contained in:
Brecht Van Lommel
2012-01-25 17:23:52 +00:00
parent 14f475fcca
commit f99343d3b8
36 changed files with 1528 additions and 243 deletions
Download Patch File Download Diff File Expand all files Collapse all files

32
intern/cycles/blender/addon/ui.py
View File

@@ -183,6 +183,38 @@ class CyclesRender_PT_layers(CyclesButtonsPanel, Panel):
layout.separator() layout.separator()
split = layout.split()
col = split.column()
col.label(text="Passes:")
col.prop(rl, "use_pass_combined")
col.prop(rl, "use_pass_z")
col.prop(rl, "use_pass_normal")
col.prop(rl, "use_pass_object_index")
col.prop(rl, "use_pass_material_index")
col.prop(rl, "use_pass_emit")
col.prop(rl, "use_pass_environment")
col = split.column()
col.label()
col.label(text="Diffuse:")
row = col.row(align=True)
row.prop(rl, "use_pass_diffuse_direct", text="Direct", toggle=True)
row.prop(rl, "use_pass_diffuse_indirect", text="Indirect", toggle=True)
row.prop(rl, "use_pass_diffuse_color", text="Color", toggle=True)
col.label(text="Glossy:")
row = col.row(align=True)
row.prop(rl, "use_pass_glossy_direct", text="Direct", toggle=True)
row.prop(rl, "use_pass_glossy_indirect", text="Indirect", toggle=True)
row.prop(rl, "use_pass_glossy_color", text="Color", toggle=True)
col.label(text="Transmission:")
row = col.row(align=True)
row.prop(rl, "use_pass_transmission_direct", text="Direct", toggle=True)
row.prop(rl, "use_pass_transmission_indirect", text="Indirect", toggle=True)
row.prop(rl, "use_pass_transmission_color", text="Color", toggle=True)
layout.separator()
rl = rd.layers[0] rl = rd.layers[0]
layout.prop(rl, "material_override", text="Material") layout.prop(rl, "material_override", text="Material")

1
intern/cycles/blender/blender_object.cpp
View File

@@ -214,6 +214,7 @@ void BlenderSync::sync_object(BL::Object b_parent, int b_index, BL::Object b_ob,
/* object sync */ /* object sync */
if(object_updated || (object->mesh && object->mesh->need_update)) { if(object_updated || (object->mesh && object->mesh->need_update)) {
object->name = b_ob.name().c_str(); object->name = b_ob.name().c_str();
object->pass_id = b_ob.pass_index();
object->tfm = tfm; object->tfm = tfm;
/* visibility flags for both parent */ /* visibility flags for both parent */

113
intern/cycles/blender/blender_session.cpp
View File

@@ -116,9 +116,68 @@ void BlenderSession::free_session()
delete session; delete session;
} }
static PassType get_pass_type(BL::RenderPass b_pass)
{
switch(b_pass.type()) {
case BL::RenderPass::type_COMBINED:
return PASS_COMBINED;
case BL::RenderPass::type_Z:
return PASS_DEPTH;
case BL::RenderPass::type_NORMAL:
return PASS_NORMAL;
case BL::RenderPass::type_OBJECT_INDEX:
return PASS_OBJECT_ID;
case BL::RenderPass::type_UV:
return PASS_UV;
case BL::RenderPass::type_MATERIAL_INDEX:
return PASS_MATERIAL_ID;
case BL::RenderPass::type_DIFFUSE_DIRECT:
return PASS_DIFFUSE_DIRECT;
case BL::RenderPass::type_GLOSSY_DIRECT:
return PASS_GLOSSY_DIRECT;
case BL::RenderPass::type_TRANSMISSION_DIRECT:
return PASS_TRANSMISSION_DIRECT;
case BL::RenderPass::type_DIFFUSE_INDIRECT:
return PASS_DIFFUSE_INDIRECT;
case BL::RenderPass::type_GLOSSY_INDIRECT:
return PASS_GLOSSY_INDIRECT;
case BL::RenderPass::type_TRANSMISSION_INDIRECT:
return PASS_TRANSMISSION_INDIRECT;
case BL::RenderPass::type_DIFFUSE_COLOR:
return PASS_DIFFUSE_COLOR;
case BL::RenderPass::type_GLOSSY_COLOR:
return PASS_GLOSSY_COLOR;
case BL::RenderPass::type_TRANSMISSION_COLOR:
return PASS_TRANSMISSION_COLOR;
case BL::RenderPass::type_EMIT:
return PASS_EMISSION;
case BL::RenderPass::type_ENVIRONMENT:
return PASS_BACKGROUND;
case BL::RenderPass::type_DIFFUSE:
case BL::RenderPass::type_SHADOW:
case BL::RenderPass::type_AO:
case BL::RenderPass::type_COLOR:
case BL::RenderPass::type_REFRACTION:
case BL::RenderPass::type_SPECULAR:
case BL::RenderPass::type_REFLECTION:
case BL::RenderPass::type_VECTOR:
case BL::RenderPass::type_MIST:
return PASS_NONE;
}
return PASS_NONE;
}
void BlenderSession::render() void BlenderSession::render()
{ {
/* get buffer parameters */ /* get buffer parameters */
SessionParams session_params = BlenderSync::get_session_params(b_userpref, b_scene, background);
BufferParams buffer_params = BlenderSync::get_buffer_params(b_scene, b_rv3d, width, height); BufferParams buffer_params = BlenderSync::get_buffer_params(b_scene, b_rv3d, width, height);
int w = buffer_params.width, h = buffer_params.height; int w = buffer_params.width, h = buffer_params.height;
@@ -143,6 +202,25 @@ void BlenderSession::render()
/* set layer */ /* set layer */
b_rlay = *b_iter; b_rlay = *b_iter;
/* add passes */
if(session_params.device.type == DEVICE_CPU) { /* todo */
BL::RenderLayer::passes_iterator b_pass_iter;
for(b_rlay.passes.begin(b_pass_iter); b_pass_iter != b_rlay.passes.end(); ++b_pass_iter) {
BL::RenderPass b_pass(*b_pass_iter);
PassType pass_type = get_pass_type(b_pass);
if(pass_type != PASS_NONE)
Pass::add(pass_type, buffer_params.passes);
}
}
scene->film->passes = buffer_params.passes;
scene->film->need_update = true;
/* update session */
session->reset(buffer_params, session_params.samples);
/* update scene */ /* update scene */
sync->sync_data(b_v3d, active); sync->sync_data(b_v3d, active);
@@ -165,22 +243,41 @@ void BlenderSession::write_render_result()
{ {
/* get state */ /* get state */
RenderBuffers *buffers = session->buffers; RenderBuffers *buffers = session->buffers;
/* copy data from device */
if(!buffers->copy_from_device())
return;
BufferParams& params = buffers->params;
float exposure = scene->film->exposure; float exposure = scene->film->exposure;
double total_time, sample_time; double total_time, sample_time;
int sample; int sample;
session->progress.get_sample(sample, total_time, sample_time); session->progress.get_sample(sample, total_time, sample_time);
/* get pixels */ vector<float> pixels(params.width*params.height*4);
float4 *pixels = buffers->copy_from_device(exposure, sample);
if(!pixels) /* copy each pass */
return; BL::RenderLayer::passes_iterator b_iter;
for(b_rlay.passes.begin(b_iter); b_iter != b_rlay.passes.end(); ++b_iter) {
BL::RenderPass b_pass(*b_iter);
/* write pixels */ /* find matching pass type */
rna_RenderLayer_rect_set(&b_rlay.ptr, (float*)pixels); PassType pass_type = get_pass_type(b_pass);
int components = b_pass.channels();
/* copy pixels */
if(buffers->get_pass(pass_type, exposure, sample, components, &pixels[0]))
rna_RenderPass_rect_set(&b_pass.ptr, &pixels[0]);
}
/* copy combined pass */
if(buffers->get_pass(PASS_COMBINED, exposure, sample, 4, &pixels[0]))
rna_RenderLayer_rect_set(&b_rlay.ptr, &pixels[0]);
/* tag result as updated */
RE_engine_update_result((RenderEngine*)b_engine.ptr.data, (RenderResult*)b_rr.ptr.data); RE_engine_update_result((RenderEngine*)b_engine.ptr.data, (RenderResult*)b_rr.ptr.data);
delete [] pixels;
} }
void BlenderSession::synchronize() void BlenderSession::synchronize()

1
intern/cycles/blender/blender_shader.cpp
View File

@@ -636,6 +636,7 @@ void BlenderSync::sync_materials()
ShaderGraph *graph = new ShaderGraph(); ShaderGraph *graph = new ShaderGraph();
shader->name = b_mat->name().c_str(); shader->name = b_mat->name().c_str();
shader->pass_id = b_mat->pass_index();
/* create nodes */ /* create nodes */
if(b_mat->use_nodes() && b_mat->node_tree()) { if(b_mat->use_nodes() && b_mat->node_tree()) {

8
intern/cycles/device/device_cpu.cpp
View File

@@ -162,7 +162,7 @@ public:
if(system_cpu_support_optimized()) { if(system_cpu_support_optimized()) {
for(int y = task.y; y < task.y + task.h; y++) { for(int y = task.y; y < task.y + task.h; y++) {
for(int x = task.x; x < task.x + task.w; x++) for(int x = task.x; x < task.x + task.w; x++)
kernel_cpu_optimized_path_trace(kg, (float4*)task.buffer, (unsigned int*)task.rng_state, kernel_cpu_optimized_path_trace(kg, (float*)task.buffer, (unsigned int*)task.rng_state,
task.sample, x, y, task.offset, task.stride); task.sample, x, y, task.offset, task.stride);
if(tasks.worker_cancel()) if(tasks.worker_cancel())
@@ -174,7 +174,7 @@ public:
{ {
for(int y = task.y; y < task.y + task.h; y++) { for(int y = task.y; y < task.y + task.h; y++) {
for(int x = task.x; x < task.x + task.w; x++) for(int x = task.x; x < task.x + task.w; x++)
kernel_cpu_path_trace(kg, (float4*)task.buffer, (unsigned int*)task.rng_state, kernel_cpu_path_trace(kg, (float*)task.buffer, (unsigned int*)task.rng_state,
task.sample, x, y, task.offset, task.stride); task.sample, x, y, task.offset, task.stride);
if(tasks.worker_cancel()) if(tasks.worker_cancel())
@@ -194,7 +194,7 @@ public:
if(system_cpu_support_optimized()) { if(system_cpu_support_optimized()) {
for(int y = task.y; y < task.y + task.h; y++) for(int y = task.y; y < task.y + task.h; y++)
for(int x = task.x; x < task.x + task.w; x++) for(int x = task.x; x < task.x + task.w; x++)
kernel_cpu_optimized_tonemap(kg, (uchar4*)task.rgba, (float4*)task.buffer, kernel_cpu_optimized_tonemap(kg, (uchar4*)task.rgba, (float*)task.buffer,
task.sample, task.resolution, x, y, task.offset, task.stride); task.sample, task.resolution, x, y, task.offset, task.stride);
} }
else else
@@ -202,7 +202,7 @@ public:
{ {
for(int y = task.y; y < task.y + task.h; y++) for(int y = task.y; y < task.y + task.h; y++)
for(int x = task.x; x < task.x + task.w; x++) for(int x = task.x; x < task.x + task.w; x++)
kernel_cpu_tonemap(kg, (uchar4*)task.rgba, (float4*)task.buffer, kernel_cpu_tonemap(kg, (uchar4*)task.rgba, (float*)task.buffer,
task.sample, task.resolution, x, y, task.offset, task.stride); task.sample, task.resolution, x, y, task.offset, task.stride);
} }
} }

2
intern/cycles/kernel/CMakeLists.txt
View File

@@ -15,6 +15,7 @@ set(SRC
set(SRC_HEADERS set(SRC_HEADERS
kernel.h kernel.h
kernel_accumulate.h
kernel_bvh.h kernel_bvh.h
kernel_camera.h kernel_camera.h
kernel_compat_cpu.h kernel_compat_cpu.h
@@ -30,6 +31,7 @@ set(SRC_HEADERS
kernel_mbvh.h kernel_mbvh.h
kernel_montecarlo.h kernel_montecarlo.h
kernel_object.h kernel_object.h
kernel_passes.h
kernel_path.h kernel_path.h
kernel_qbvh.h kernel_qbvh.h
kernel_random.h kernel_random.h

4
intern/cycles/kernel/kernel.cl
View File

@@ -28,7 +28,7 @@
__kernel void kernel_ocl_path_trace( __kernel void kernel_ocl_path_trace(
__constant KernelData *data, __constant KernelData *data,
__global float4 *buffer, __global float *buffer,
__global uint *rng_state, __global uint *rng_state,
#define KERNEL_TEX(type, ttype, name) \ #define KERNEL_TEX(type, ttype, name) \
@@ -56,7 +56,7 @@ __kernel void kernel_ocl_path_trace(
__kernel void kernel_ocl_tonemap( __kernel void kernel_ocl_tonemap(
__constant KernelData *data, __constant KernelData *data,
__global uchar4 *rgba, __global uchar4 *rgba,
__global float4 *buffer, __global float *buffer,
#define KERNEL_TEX(type, ttype, name) \ #define KERNEL_TEX(type, ttype, name) \
__global type *name, __global type *name,

4
intern/cycles/kernel/kernel.cpp
View File

@@ -204,14 +204,14 @@ void kernel_tex_copy(KernelGlobals *kg, const char *name, device_ptr mem, size_t
/* Path Tracing */ /* Path Tracing */
void kernel_cpu_path_trace(KernelGlobals *kg, float4 *buffer, unsigned int *rng_state, int sample, int x, int y, int offset, int stride) void kernel_cpu_path_trace(KernelGlobals *kg, float *buffer, unsigned int *rng_state, int sample, int x, int y, int offset, int stride)
{ {
kernel_path_trace(kg, buffer, rng_state, sample, x, y, offset, stride); kernel_path_trace(kg, buffer, rng_state, sample, x, y, offset, stride);
} }
/* Tonemapping */ /* Tonemapping */
void kernel_cpu_tonemap(KernelGlobals *kg, uchar4 *rgba, float4 *buffer, int sample, int resolution, int x, int y, int offset, int stride) void kernel_cpu_tonemap(KernelGlobals *kg, uchar4 *rgba, float *buffer, int sample, int resolution, int x, int y, int offset, int stride)
{ {
kernel_film_tonemap(kg, rgba, buffer, sample, resolution, x, y, offset, stride); kernel_film_tonemap(kg, rgba, buffer, sample, resolution, x, y, offset, stride);
} }

4
intern/cycles/kernel/kernel.cu
View File

@@ -26,7 +26,7 @@
#include "kernel_path.h" #include "kernel_path.h"
#include "kernel_displace.h" #include "kernel_displace.h"
extern "C" __global__ void kernel_cuda_path_trace(float4 *buffer, uint *rng_state, int sample, int sx, int sy, int sw, int sh, int offset, int stride) extern "C" __global__ void kernel_cuda_path_trace(float *buffer, uint *rng_state, int sample, int sx, int sy, int sw, int sh, int offset, int stride)
{ {
int x = sx + blockDim.x*blockIdx.x + threadIdx.x; int x = sx + blockDim.x*blockIdx.x + threadIdx.x;
int y = sy + blockDim.y*blockIdx.y + threadIdx.y; int y = sy + blockDim.y*blockIdx.y + threadIdx.y;
@@ -35,7 +35,7 @@ extern "C" __global__ void kernel_cuda_path_trace(float4 *buffer, uint *rng_stat
kernel_path_trace(NULL, buffer, rng_state, sample, x, y, offset, stride); kernel_path_trace(NULL, buffer, rng_state, sample, x, y, offset, stride);
} }
extern "C" __global__ void kernel_cuda_tonemap(uchar4 *rgba, float4 *buffer, int sample, int resolution, int sx, int sy, int sw, int sh, int offset, int stride) extern "C" __global__ void kernel_cuda_tonemap(uchar4 *rgba, float *buffer, int sample, int resolution, int sx, int sy, int sw, int sh, int offset, int stride)
{ {
int x = sx + blockDim.x*blockIdx.x + threadIdx.x; int x = sx + blockDim.x*blockIdx.x + threadIdx.x;
int y = sy + blockDim.y*blockIdx.y + threadIdx.y; int y = sy + blockDim.y*blockIdx.y + threadIdx.y;

8
intern/cycles/kernel/kernel.h
View File

@@ -36,17 +36,17 @@ bool kernel_osl_use(KernelGlobals *kg);
void kernel_const_copy(KernelGlobals *kg, const char *name, void *host, size_t size); void kernel_const_copy(KernelGlobals *kg, const char *name, void *host, size_t size);
void kernel_tex_copy(KernelGlobals *kg, const char *name, device_ptr mem, size_t width, size_t height); void kernel_tex_copy(KernelGlobals *kg, const char *name, device_ptr mem, size_t width, size_t height);
void kernel_cpu_path_trace(KernelGlobals *kg, float4 *buffer, unsigned int *rng_state, void kernel_cpu_path_trace(KernelGlobals *kg, float *buffer, unsigned int *rng_state,
int sample, int x, int y, int offset, int stride); int sample, int x, int y, int offset, int stride);
void kernel_cpu_tonemap(KernelGlobals *kg, uchar4 *rgba, float4 *buffer, void kernel_cpu_tonemap(KernelGlobals *kg, uchar4 *rgba, float *buffer,
int sample, int resolution, int x, int y, int offset, int stride); int sample, int resolution, int x, int y, int offset, int stride);
void kernel_cpu_shader(KernelGlobals *kg, uint4 *input, float4 *output, void kernel_cpu_shader(KernelGlobals *kg, uint4 *input, float4 *output,
int type, int i); int type, int i);
#ifdef WITH_OPTIMIZED_KERNEL #ifdef WITH_OPTIMIZED_KERNEL
void kernel_cpu_optimized_path_trace(KernelGlobals *kg, float4 *buffer, unsigned int *rng_state, void kernel_cpu_optimized_path_trace(KernelGlobals *kg, float *buffer, unsigned int *rng_state,
int sample, int x, int y, int offset, int stride); int sample, int x, int y, int offset, int stride);
void kernel_cpu_optimized_tonemap(KernelGlobals *kg, uchar4 *rgba, float4 *buffer, void kernel_cpu_optimized_tonemap(KernelGlobals *kg, uchar4 *rgba, float *buffer,
int sample, int resolution, int x, int y, int offset, int stride); int sample, int resolution, int x, int y, int offset, int stride);
void kernel_cpu_optimized_shader(KernelGlobals *kg, uint4 *input, float4 *output, void kernel_cpu_optimized_shader(KernelGlobals *kg, uint4 *input, float4 *output,
int type, int i); int type, int i);

283
intern/cycles/kernel/kernel_accumulate.h Normal file
View File

@@ -0,0 +1,283 @@
/*
* Copyright 2011, Blender Foundation.
*
* This program is free software; you can redistribute it and/or
* modify it under the terms of the GNU General Public License
* as published by the Free Software Foundation; either version 2
* of the License, or (at your option) any later version.
*
* This program is distributed in the hope that it will be useful,
* but WITHOUT ANY WARRANTY; without even the implied warranty of
* MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
* GNU General Public License for more details.
*
* You should have received a copy of the GNU General Public License
* along with this program; if not, write to the Free Software Foundation,
* Inc., 51 Franklin Street, Fifth Floor, Boston, MA 02110-1301, USA.
*/
CCL_NAMESPACE_BEGIN
/* BSDF Eval
*
* BSDF evaluation result, split per BSDF type. This is used to accumulate
* render passes separately. */
__device_inline void bsdf_eval_init(BsdfEval *eval, ClosureType type, float3 value, int use_light_pass)
{
#ifdef __PASSES__
eval->use_light_pass = use_light_pass;
if(eval->use_light_pass) {
eval->diffuse = make_float3(0.0f, 0.0f, 0.0f);
eval->glossy = make_float3(0.0f, 0.0f, 0.0f);
eval->transmission = make_float3(0.0f, 0.0f, 0.0f);
eval->transparent = make_float3(0.0f, 0.0f, 0.0f);
if(type == CLOSURE_BSDF_TRANSPARENT_ID)
eval->transparent = value;
else if(CLOSURE_IS_BSDF_DIFFUSE(type))
eval->diffuse = value;
else if(CLOSURE_IS_BSDF_GLOSSY(type))
eval->glossy = value;
else
eval->transmission = value;
}
else
eval->diffuse = value;
#else
*eval = value;
#endif
}
__device_inline void bsdf_eval_accum(BsdfEval *eval, ClosureType type, float3 value)
{
#ifdef __PASSES__
if(eval->use_light_pass) {
if(CLOSURE_IS_BSDF_DIFFUSE(type))
eval->diffuse += value;
else if(CLOSURE_IS_BSDF_GLOSSY(type))
eval->glossy += value;
else
eval->transmission += value;
/* skipping transparent, this function is used by for eval(), will be zero then */
}
else
eval->diffuse += value;
#else
*eval += value;
#endif
}
__device_inline bool bsdf_eval_is_zero(BsdfEval *eval)
{
#ifdef __PASSES__
if(eval->use_light_pass) {
return is_zero(eval->diffuse)
&& is_zero(eval->glossy)
&& is_zero(eval->transmission)
&& is_zero(eval->transparent);
}
else
return is_zero(eval->diffuse);
#else
return is_zero(*eval);
#endif
}
__device_inline void bsdf_eval_mul(BsdfEval *eval, float3 value)
{
#ifdef __PASSES__
if(eval->use_light_pass) {
eval->diffuse *= value;
eval->glossy *= value;
eval->transmission *= value;
/* skipping transparent, this function is used by for eval(), will be zero then */
}
else
eval->diffuse *= value;
#else
*eval *= value;
#endif
}
/* Path Radiance
*
* We accumulate different render passes separately. After summing at the end
* to get the combined result, it should be identical. We definte directly
* visible as the first non-transparent hit, while indirectly visible are the
* bounces after that. */
__device_inline void path_radiance_init(PathRadiance *L, int use_light_pass)
{
/* clear all */
#ifdef __PASSES__
L->use_light_pass = use_light_pass;
if(use_light_pass) {
L->indirect = make_float3(0.0f, 0.0f, 0.0f);
L->direct_throughput = make_float3(0.0f, 0.0f, 0.0f);
L->direct_emission = make_float3(0.0f, 0.0f, 0.0f);
L->color_diffuse = make_float3(0.0f, 0.0f, 0.0f);
L->color_glossy = make_float3(0.0f, 0.0f, 0.0f);
L->color_transmission = make_float3(0.0f, 0.0f, 0.0f);
L->direct_diffuse = make_float3(0.0f, 0.0f, 0.0f);
L->direct_glossy = make_float3(0.0f, 0.0f, 0.0f);
L->direct_transmission = make_float3(0.0f, 0.0f, 0.0f);
L->indirect_diffuse = make_float3(0.0f, 0.0f, 0.0f);
L->indirect_glossy = make_float3(0.0f, 0.0f, 0.0f);
L->indirect_transmission = make_float3(0.0f, 0.0f, 0.0f);
L->emission = make_float3(0.0f, 0.0f, 0.0f);
L->background = make_float3(0.0f, 0.0f, 0.0f);
}
else
L->emission = make_float3(0.0f, 0.0f, 0.0f);
#else
*L = make_float3(0.0f, 0.0f, 0.0f);
#endif
}
__device_inline void path_radiance_bsdf_bounce(PathRadiance *L, float3 *throughput,
BsdfEval *bsdf_eval, float bsdf_pdf, int bounce, int bsdf_label)
{
float inverse_pdf = 1.0f/bsdf_pdf;
#ifdef __PASSES__
if(L->use_light_pass) {
if(bounce == 0) {
if(bsdf_label & LABEL_TRANSPARENT) {
/* transparent bounce before first hit */
*throughput *= bsdf_eval->transparent*inverse_pdf;
}
else {
/* first on directly visible surface */
float3 value = *throughput*inverse_pdf;
L->indirect_diffuse = bsdf_eval->diffuse*value;
L->indirect_glossy = bsdf_eval->glossy*value;
L->indirect_transmission = bsdf_eval->transmission*value;
*throughput = L->indirect_diffuse + L->indirect_glossy + L->indirect_transmission;
L->direct_throughput = *throughput;
}
}
else {
/* indirectly visible through BSDF */
float3 sum = (bsdf_eval->diffuse + bsdf_eval->glossy + bsdf_eval->transmission + bsdf_eval->transparent)*inverse_pdf;
*throughput *= sum;
}
}
else
*throughput *= bsdf_eval->diffuse*inverse_pdf;
#else
*throughput *= *bsdf_eval*inverse_pdf;
#endif
}
__device_inline void path_radiance_accum_emission(PathRadiance *L, float3 throughput, float3 value, int bounce)
{
#ifdef __PASSES__
if(L->use_light_pass) {
if(bounce == 0)
L->emission += throughput*value;
else if(bounce == 1)
L->direct_emission += throughput*value;
else
L->indirect += throughput*value;
}
else
L->emission += throughput*value;
#else
*L += throughput*value;
#endif
}
__device_inline void path_radiance_accum_light(PathRadiance *L, float3 throughput, BsdfEval *bsdf_eval, int bounce)
{
#ifdef __PASSES__
if(L->use_light_pass) {
if(bounce == 0) {
/* directly visible lighting */
L->direct_diffuse += throughput*bsdf_eval->diffuse;
L->direct_glossy += throughput*bsdf_eval->glossy;
L->direct_transmission += throughput*bsdf_eval->transmission;
}
else {
/* indirectly visible lighting after BSDF bounce */
float3 sum = bsdf_eval->diffuse + bsdf_eval->glossy + bsdf_eval->transmission;
L->indirect += throughput*sum;
}
}
else
L->emission += throughput*bsdf_eval->diffuse;
#else
*L += throughput*(*bsdf_eval);
#endif
}
__device_inline void path_radiance_accum_background(PathRadiance *L, float3 throughput, float3 value, int bounce)
{
#ifdef __PASSES__
if(L->use_light_pass) {
if(bounce == 0)
L->background += throughput*value;
else if(bounce == 1)
L->direct_emission += throughput*value;
else
L->indirect += throughput*value;
}
else
L->emission += throughput*value;
#else
*L += throughput*value;
#endif
}
__device_inline float3 safe_divide_color(float3 a, float3 b)
{
float x, y, z;
x = (b.x != 0.0f)? a.x/b.x: 0.0f;
y = (b.y != 0.0f)? a.y/b.y: 0.0f;
z = (b.z != 0.0f)? a.z/b.z: 0.0f;
return make_float3(x, y, z);
}
__device_inline float3 path_radiance_sum(PathRadiance *L)
{
#ifdef __PASSES__
if(L->use_light_pass) {
/* this division is a bit ugly, but means we only have to keep track of
only a single throughput further along the path, here we recover just
the indirect parth that is not influenced by any particular BSDF type */
L->direct_emission = safe_divide_color(L->direct_emission, L->direct_throughput);
L->direct_diffuse += L->indirect_diffuse*L->direct_emission;
L->direct_glossy += L->indirect_glossy*L->direct_emission;
L->direct_transmission += L->indirect_transmission*L->direct_emission;
L->indirect = safe_divide_color(L->indirect, L->direct_throughput);
L->indirect_diffuse *= L->indirect;
L->indirect_glossy *= L->indirect;
L->indirect_transmission *= L->indirect;
return L->emission + L->background
+ L->direct_diffuse + L->direct_glossy + L->direct_transmission
+ L->indirect_diffuse + L->indirect_glossy + L->indirect_transmission;
}
else
return L->emission;
#else
return *L;
#endif
}
CCL_NAMESPACE_END

23
intern/cycles/kernel/kernel_emission.h
View File

@@ -57,7 +57,7 @@ __device float3 direct_emissive_eval(KernelGlobals *kg, float rando,
} }
__device bool direct_emission(KernelGlobals *kg, ShaderData *sd, int lindex, __device bool direct_emission(KernelGlobals *kg, ShaderData *sd, int lindex,
float randt, float rando, float randu, float randv, Ray *ray, float3 *eval) float randt, float rando, float randu, float randv, Ray *ray, BsdfEval *eval)
{ {
LightSample ls; LightSample ls;
@@ -83,32 +83,33 @@ __device bool direct_emission(KernelGlobals *kg, ShaderData *sd, int lindex,
return false; return false;
/* evaluate closure */ /* evaluate closure */
*eval = direct_emissive_eval(kg, rando, &ls, randu, randv, -ls.D); float3 light_eval = direct_emissive_eval(kg, rando, &ls, randu, randv, -ls.D);
if(is_zero(*eval)) if(is_zero(light_eval))
return false; return false;
/* todo: use visbility flag to skip lights */ /* todo: use visbility flag to skip lights */
/* evaluate BSDF at shading point */ /* evaluate BSDF at shading point */
float bsdf_pdf; float bsdf_pdf;
float3 bsdf_eval = shader_bsdf_eval(kg, sd, ls.D, &bsdf_pdf);
*eval *= bsdf_eval/pdf; shader_bsdf_eval(kg, sd, ls.D, eval, &bsdf_pdf);
if(is_zero(*eval))
return false;
if(ls.prim != ~0 || ls.type == LIGHT_BACKGROUND) { if(ls.prim != ~0 || ls.type == LIGHT_BACKGROUND) {
/* multiple importance sampling */ /* multiple importance sampling */
float mis_weight = power_heuristic(pdf, bsdf_pdf); float mis_weight = power_heuristic(pdf, bsdf_pdf);
*eval *= mis_weight; light_eval *= mis_weight;
} }
/* todo: clean up these weights */ /* todo: clean up these weights */
else if(ls.shader & SHADER_AREA_LIGHT) else if(ls.shader & SHADER_AREA_LIGHT)
*eval *= 0.25f; /* area lamp */ light_eval *= 0.25f; /* area lamp */
else if(ls.t != FLT_MAX) else if(ls.t != FLT_MAX)
*eval *= 0.25f*M_1_PI_F; /* point lamp */ light_eval *= 0.25f*M_1_PI_F; /* point lamp */
bsdf_eval_mul(eval, light_eval/pdf);
if(bsdf_eval_is_zero(eval))
return false;
if(ls.shader & SHADER_CAST_SHADOW) { if(ls.shader & SHADER_CAST_SHADOW) {
/* setup ray */ /* setup ray */

13
intern/cycles/kernel/kernel_film.h
View File

@@ -48,15 +48,22 @@ __device uchar4 film_float_to_byte(float4 color)
return result; return result;
} }
__device void kernel_film_tonemap(KernelGlobals *kg, __global uchar4 *rgba, __global float4 *buffer, int sample, int resolution, int x, int y, int offset, int stride) __device void kernel_film_tonemap(KernelGlobals *kg,
__global uchar4 *rgba, __global float *buffer,
int sample, int resolution, int x, int y, int offset, int stride)
{ {
/* buffer offset */
int index = offset + x + y*stride; int index = offset + x + y*stride;
float4 irradiance = buffer[index];
rgba += index;
buffer += index*kernel_data.film.pass_stride;
/* map colors */
float4 irradiance = *(float4*)buffer;
float4 float_result = film_map(kg, irradiance, sample); float4 float_result = film_map(kg, irradiance, sample);
uchar4 byte_result = film_float_to_byte(float_result); uchar4 byte_result = film_float_to_byte(float_result);
rgba[index] = byte_result; *rgba = byte_result;
} }
CCL_NAMESPACE_END CCL_NAMESPACE_END

10
intern/cycles/kernel/kernel_object.h
View File

@@ -64,5 +64,15 @@ __device_inline float object_surface_area(KernelGlobals *kg, int object)
return f.x; return f.x;
} }
__device_inline float object_pass_id(KernelGlobals *kg, int object)
{
if(object == ~0)
return 0.0f;
int offset = object*OBJECT_SIZE + OBJECT_PROPERTIES;
float4 f = kernel_tex_fetch(__objects, offset);
return f.y;
}
CCL_NAMESPACE_END CCL_NAMESPACE_END

4
intern/cycles/kernel/kernel_optimized.cpp
View File

@@ -35,14 +35,14 @@ CCL_NAMESPACE_BEGIN
/* Path Tracing */ /* Path Tracing */
void kernel_cpu_optimized_path_trace(KernelGlobals *kg, float4 *buffer, unsigned int *rng_state, int sample, int x, int y, int offset, int stride) void kernel_cpu_optimized_path_trace(KernelGlobals *kg, float *buffer, unsigned int *rng_state, int sample, int x, int y, int offset, int stride)
{ {
kernel_path_trace(kg, buffer, rng_state, sample, x, y, offset, stride); kernel_path_trace(kg, buffer, rng_state, sample, x, y, offset, stride);
} }
/* Tonemapping */ /* Tonemapping */
void kernel_cpu_optimized_tonemap(KernelGlobals *kg, uchar4 *rgba, float4 *buffer, int sample, int resolution, int x, int y, int offset, int stride) void kernel_cpu_optimized_tonemap(KernelGlobals *kg, uchar4 *rgba, float *buffer, int sample, int resolution, int x, int y, int offset, int stride)
{ {
kernel_film_tonemap(kg, rgba, buffer, sample, resolution, x, y, offset, stride); kernel_film_tonemap(kg, rgba, buffer, sample, resolution, x, y, offset, stride);
} }

146
intern/cycles/kernel/kernel_passes.h Normal file
View File

@@ -0,0 +1,146 @@
/*
* Copyright 2011, Blender Foundation.
*
* This program is free software; you can redistribute it and/or
* modify it under the terms of the GNU General Public License
* as published by the Free Software Foundation; either version 2
* of the License, or (at your option) any later version.
*
* This program is distributed in the hope that it will be useful,
* but WITHOUT ANY WARRANTY; without even the implied warranty of
* MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
* GNU General Public License for more details.
*
* You should have received a copy of the GNU General Public License
* along with this program; if not, write to the Free Software Foundation,
* Inc., 51 Franklin Street, Fifth Floor, Boston, MA 02110-1301, USA.
*/
CCL_NAMESPACE_BEGIN
__device_inline void kernel_write_pass_float(__global float *buffer, int sample, float value)
{
float *buf = buffer;
*buf = (sample == 0)? value: *buf + value;
}
__device_inline void kernel_write_pass_float3(__global float *buffer, int sample, float3 value)
{
float3 *buf = (float3*)buffer;
*buf = (sample == 0)? value: *buf + value;
}
__device_inline void kernel_write_pass_float4(__global float *buffer, int sample, float4 value)
{
float4 *buf = (float4*)buffer;
*buf = (sample == 0)? value: *buf + value;
}
__device_inline void kernel_clear_passes(__global float *buffer, int sample, int pass_stride)
{
#ifdef __PASSES__
if(sample == 0 && pass_stride != 4)
for(int i = 4; i < pass_stride; i++)
buffer[i] = 0.0f;
#endif
}
__device void kernel_write_data_passes(KernelGlobals *kg, __global float *buffer, PathRadiance *L,
ShaderData *sd, int sample, int path_flag, float3 throughput)
{
#ifdef __PASSES__
if(!(path_flag & PATH_RAY_CAMERA))
return;
int flag = kernel_data.film.pass_flag;
if(!(flag & PASS_ALL))
return;
/* todo: add alpha treshold */
if(!(path_flag & PATH_RAY_TRANSPARENT)) {
if(sample == 0) {
if(flag & PASS_DEPTH) {
Transform tfm = kernel_data.cam.worldtocamera;
float depth = len(transform(&tfm, sd->P));
kernel_write_pass_float(buffer + kernel_data.film.pass_depth, sample, depth);
}
if(flag & PASS_OBJECT_ID) {
float id = object_pass_id(kg, sd->object);
kernel_write_pass_float(buffer + kernel_data.film.pass_object_id, sample, id);
}
if(flag & PASS_MATERIAL_ID) {
float id = shader_pass_id(kg, sd);
kernel_write_pass_float(buffer + kernel_data.film.pass_material_id, sample, id);
}
}
if(flag & PASS_NORMAL) {
float3 normal = sd->N;
kernel_write_pass_float3(buffer + kernel_data.film.pass_normal, sample, normal);
}
if(flag & PASS_UV) {
float3 uv = make_float3(0.0f, 0.0f, 0.0f); /* todo: request and lookup */
kernel_write_pass_float3(buffer + kernel_data.film.pass_uv, sample, uv);
}
}
if(flag & (PASS_DIFFUSE_INDIRECT|PASS_DIFFUSE_COLOR|PASS_DIFFUSE_DIRECT))
L->color_diffuse += shader_bsdf_diffuse(kg, sd)*throughput;
if(flag & (PASS_GLOSSY_INDIRECT|PASS_GLOSSY_COLOR|PASS_GLOSSY_DIRECT))
L->color_glossy += shader_bsdf_glossy(kg, sd)*throughput;
if(flag & (PASS_TRANSMISSION_INDIRECT|PASS_TRANSMISSION_COLOR|PASS_TRANSMISSION_DIRECT))
L->color_transmission += shader_bsdf_transmission(kg, sd)*throughput;
#endif
}
__device void kernel_write_light_passes(KernelGlobals *kg, __global float *buffer, PathRadiance *L, int sample)
{
#ifdef __PASSES__
int flag = kernel_data.film.pass_flag;
if(!kernel_data.film.use_light_pass)
return;
if(flag & PASS_DIFFUSE_INDIRECT) {
float3 color = safe_divide_color(L->indirect_diffuse, L->color_diffuse);
kernel_write_pass_float3(buffer + kernel_data.film.pass_diffuse_indirect, sample, color);
}
if(flag & PASS_GLOSSY_INDIRECT) {
float3 color = safe_divide_color(L->indirect_glossy, L->color_glossy);
kernel_write_pass_float3(buffer + kernel_data.film.pass_glossy_indirect, sample, color);
}
if(flag & PASS_TRANSMISSION_INDIRECT) {
float3 color = safe_divide_color(L->indirect_transmission, L->color_transmission);
kernel_write_pass_float3(buffer + kernel_data.film.pass_transmission_indirect, sample, color);
}
if(flag & PASS_DIFFUSE_DIRECT) {
float3 color = safe_divide_color(L->direct_diffuse, L->color_diffuse);
kernel_write_pass_float3(buffer + kernel_data.film.pass_diffuse_direct, sample, color);
}
if(flag & PASS_GLOSSY_DIRECT) {
float3 color = safe_divide_color(L->direct_glossy, L->color_glossy);
kernel_write_pass_float3(buffer + kernel_data.film.pass_glossy_direct, sample, color);
}
if(flag & PASS_TRANSMISSION_DIRECT) {
float3 color = safe_divide_color(L->direct_transmission, L->color_transmission);
kernel_write_pass_float3(buffer + kernel_data.film.pass_transmission_direct, sample, color);
}
if(flag & PASS_EMISSION)
kernel_write_pass_float3(buffer + kernel_data.film.pass_emission, sample, L->emission);
if(flag & PASS_BACKGROUND)
kernel_write_pass_float3(buffer + kernel_data.film.pass_background, sample, L->background);
if(flag & PASS_DIFFUSE_COLOR)
kernel_write_pass_float3(buffer + kernel_data.film.pass_diffuse_color, sample, L->color_diffuse);
if(flag & PASS_GLOSSY_COLOR)
kernel_write_pass_float3(buffer + kernel_data.film.pass_glossy_color, sample, L->color_glossy);
if(flag & PASS_TRANSMISSION_COLOR)
kernel_write_pass_float3(buffer + kernel_data.film.pass_transmission_color, sample, L->color_transmission);
#endif
}
CCL_NAMESPACE_END

109
intern/cycles/kernel/kernel_path.h
View File

@@ -25,39 +25,16 @@
#else #else
#include "kernel_bvh.h" #include "kernel_bvh.h"
#endif #endif
#include "kernel_accumulate.h"
#include "kernel_camera.h" #include "kernel_camera.h"
#include "kernel_shader.h" #include "kernel_shader.h"
#include "kernel_light.h" #include "kernel_light.h"
#include "kernel_emission.h" #include "kernel_emission.h"
#include "kernel_random.h" #include "kernel_random.h"
#include "kernel_passes.h"
CCL_NAMESPACE_BEGIN CCL_NAMESPACE_BEGIN
#ifdef __MODIFY_TP__
__device float3 path_terminate_modified_throughput(KernelGlobals *kg, __global float3 *buffer, int x, int y, int offset, int stride, int sample)
{
/* modify throughput to influence path termination probability, to avoid
darker regions receiving fewer samples than lighter regions. also RGB
are weighted differently. proper validation still remains to be done. */
const float3 weights = make_float3(1.0f, 1.33f, 0.66f);
const float3 one = make_float3(1.0f, 1.0f, 1.0f);
const int minsample = 5;
const float minL = 0.1f;
if(sample >= minsample) {
float3 L = buffer[offset + x + y*stride];
float3 Lmin = make_float3(minL, minL, minL);
float correct = (float)(sample+1)/(float)sample;
L = film_map(L*correct, sample);
return weights/clamp(L, Lmin, one);
}
return weights;
}
#endif
typedef struct PathState { typedef struct PathState {
uint flag; uint flag;
int bounce; int bounce;
@@ -168,7 +145,7 @@ __device_inline float path_state_terminate_probability(KernelGlobals *kg, PathSt
return average(throughput); return average(throughput);
} }
__device_inline bool shadow_blocked(KernelGlobals *kg, PathState *state, Ray *ray, Intersection *isect, float3 *light_L) __device_inline bool shadow_blocked(KernelGlobals *kg, PathState *state, Ray *ray, Intersection *isect, BsdfEval *L_light)
{ {
if(ray->t == 0.0f) if(ray->t == 0.0f)
return false; return false;
@@ -208,7 +185,7 @@ __device_inline bool shadow_blocked(KernelGlobals *kg, PathState *state, Ray *ra
} }
if(!scene_intersect(kg, ray, PATH_RAY_SHADOW_TRANSPARENT, isect)) { if(!scene_intersect(kg, ray, PATH_RAY_SHADOW_TRANSPARENT, isect)) {
*light_L *= throughput; bsdf_eval_mul(L_light, throughput);
return false; return false;
} }
@@ -234,11 +211,14 @@ __device_inline bool shadow_blocked(KernelGlobals *kg, PathState *state, Ray *ra
return result; return result;
} }
__device float4 kernel_path_integrate(KernelGlobals *kg, RNG *rng, int sample, Ray ray, float3 throughput) __device float4 kernel_path_integrate(KernelGlobals *kg, RNG *rng, int sample, Ray ray, __global float *buffer)
{ {
/* initialize */ /* initialize */
float3 L = make_float3(0.0f, 0.0f, 0.0f); PathRadiance L;
float Ltransparent = 0.0f; float3 throughput = make_float3(1.0f, 1.0f, 1.0f);
float L_transparent = 0.0f;
path_radiance_init(&L, kernel_data.film.use_light_pass);
#ifdef __EMISSION__ #ifdef __EMISSION__
float ray_pdf = 0.0f; float ray_pdf = 0.0f;
@@ -257,12 +237,12 @@ __device float4 kernel_path_integrate(KernelGlobals *kg, RNG *rng, int sample, R
if(!scene_intersect(kg, &ray, visibility, &isect)) { if(!scene_intersect(kg, &ray, visibility, &isect)) {
/* eval background shader if nothing hit */ /* eval background shader if nothing hit */
if(kernel_data.background.transparent && (state.flag & PATH_RAY_CAMERA)) { if(kernel_data.background.transparent && (state.flag & PATH_RAY_CAMERA)) {
Ltransparent += average(throughput); L_transparent += average(throughput);
} }
else { else {
/* sample background shader */ /* sample background shader */
float3 background_L = indirect_background(kg, &ray, state.flag, ray_pdf); float3 L_background = indirect_background(kg, &ray, state.flag, ray_pdf);
L += throughput*background_L; path_radiance_accum_background(&L, throughput, L_background, state.bounce);
} }
break; break;
@@ -274,19 +254,24 @@ __device float4 kernel_path_integrate(KernelGlobals *kg, RNG *rng, int sample, R
float rbsdf = path_rng(kg, rng, sample, rng_offset + PRNG_BSDF); float rbsdf = path_rng(kg, rng, sample, rng_offset + PRNG_BSDF);
shader_eval_surface(kg, &sd, rbsdf, state.flag); shader_eval_surface(kg, &sd, rbsdf, state.flag);
kernel_write_data_passes(kg, buffer, &L, &sd, sample, state.flag, throughput);
#ifdef __HOLDOUT__ #ifdef __HOLDOUT__
if((sd.flag & SD_HOLDOUT) && (state.flag & PATH_RAY_CAMERA)) { if((sd.flag & SD_HOLDOUT) && (state.flag & PATH_RAY_CAMERA)) {
float3 holdout_weight = shader_holdout_eval(kg, &sd); float3 holdout_weight = shader_holdout_eval(kg, &sd);
if(kernel_data.background.transparent) if(kernel_data.background.transparent)
Ltransparent += average(holdout_weight*throughput); /* any throughput is ok, should all be identical here */
L_transparent += average(holdout_weight*throughput);
} }
#endif #endif
#ifdef __EMISSION__ #ifdef __EMISSION__
/* emission */ /* emission */
if(sd.flag & SD_EMISSION) if(sd.flag & SD_EMISSION) {
L += throughput*indirect_emission(kg, &sd, isect.t, state.flag, ray_pdf); float3 emission = indirect_emission(kg, &sd, isect.t, state.flag, ray_pdf);
path_radiance_accum_emission(&L, throughput, emission, state.bounce);
}
#endif #endif
/* path termination. this is a strange place to put the termination, it's /* path termination. this is a strange place to put the termination, it's
@@ -310,7 +295,7 @@ __device float4 kernel_path_integrate(KernelGlobals *kg, RNG *rng, int sample, R
float light_v = path_rng(kg, rng, sample, rng_offset + PRNG_LIGHT_V); float light_v = path_rng(kg, rng, sample, rng_offset + PRNG_LIGHT_V);
Ray light_ray; Ray light_ray;
float3 light_L; BsdfEval L_light;
#ifdef __MULTI_LIGHT__ #ifdef __MULTI_LIGHT__
/* index -1 means randomly sample from distribution */ /* index -1 means randomly sample from distribution */
@@ -320,10 +305,10 @@ __device float4 kernel_path_integrate(KernelGlobals *kg, RNG *rng, int sample, R
#else #else
const int i = -1; const int i = -1;
#endif #endif
if(direct_emission(kg, &sd, i, light_t, light_o, light_u, light_v, &light_ray, &light_L)) { if(direct_emission(kg, &sd, i, light_t, light_o, light_u, light_v, &light_ray, &L_light)) {
/* trace shadow ray */ /* trace shadow ray */
if(!shadow_blocked(kg, &state, &light_ray, &isect, &light_L)) if(!shadow_blocked(kg, &state, &light_ray, &isect, &L_light))
L += throughput*light_L; path_radiance_accum_light(&L, throughput, &L_light, state.bounce);
} }
#ifdef __MULTI_LIGHT__ #ifdef __MULTI_LIGHT__
} }
@@ -338,7 +323,7 @@ __device float4 kernel_path_integrate(KernelGlobals *kg, RNG *rng, int sample, R
/* sample BSDF */ /* sample BSDF */
float bsdf_pdf; float bsdf_pdf;
float3 bsdf_eval; BsdfEval bsdf_eval;
float3 bsdf_omega_in; float3 bsdf_omega_in;
differential3 bsdf_domega_in; differential3 bsdf_domega_in;
float bsdf_u = path_rng(kg, rng, sample, rng_offset + PRNG_BSDF_U); float bsdf_u = path_rng(kg, rng, sample, rng_offset + PRNG_BSDF_U);
@@ -350,11 +335,11 @@ __device float4 kernel_path_integrate(KernelGlobals *kg, RNG *rng, int sample, R
shader_release(kg, &sd); shader_release(kg, &sd);
if(bsdf_pdf == 0.0f || is_zero(bsdf_eval)) if(bsdf_pdf == 0.0f || bsdf_eval_is_zero(&bsdf_eval))
break; break;
/* modify throughput */ /* modify throughput */
throughput *= bsdf_eval/bsdf_pdf; path_radiance_bsdf_bounce(&L, &throughput, &bsdf_eval, bsdf_pdf, state.bounce, label);
/* set labels */ /* set labels */
#if defined(__EMISSION__) || defined(__BACKGROUND__) #if defined(__EMISSION__) || defined(__BACKGROUND__)
@@ -374,18 +359,33 @@ __device float4 kernel_path_integrate(KernelGlobals *kg, RNG *rng, int sample, R
#endif #endif
} }
return make_float4(L.x, L.y, L.z, 1.0f - Ltransparent); float3 L_sum = path_radiance_sum(&L);
kernel_write_light_passes(kg, buffer, &L, sample);
return make_float4(L_sum.x, L_sum.y, L_sum.z, 1.0f - L_transparent);
} }
__device void kernel_path_trace(KernelGlobals *kg, __global float4 *buffer, __global uint *rng_state, int sample, int x, int y, int offset, int stride) __device void kernel_path_trace(KernelGlobals *kg,
__global float *buffer, __global uint *rng_state,
int sample, int x, int y, int offset, int stride)
{ {
/* buffer offset */
int index = offset + x + y*stride;
int pass_stride = kernel_data.film.pass_stride;
rng_state += index;
buffer += index*pass_stride;
kernel_clear_passes(buffer, sample, pass_stride);
/* initialize random numbers */ /* initialize random numbers */
RNG rng; RNG rng;
float filter_u; float filter_u;
float filter_v; float filter_v;
path_rng_init(kg, rng_state, sample, &rng, x, y, offset, stride, &filter_u, &filter_v); path_rng_init(kg, rng_state, sample, &rng, x, y, &filter_u, &filter_v);
/* sample camera ray */ /* sample camera ray */
Ray ray; Ray ray;
@@ -396,23 +396,12 @@ __device void kernel_path_trace(KernelGlobals *kg, __global float4 *buffer, __gl
camera_sample(kg, x, y, filter_u, filter_v, lens_u, lens_v, &ray); camera_sample(kg, x, y, filter_u, filter_v, lens_u, lens_v, &ray);
/* integrate */ /* integrate */
#ifdef __MODIFY_TP__ float4 L = kernel_path_integrate(kg, &rng, sample, ray, buffer);
float3 throughput = path_terminate_modified_throughput(kg, buffer, x, y, offset, stride, sample);
float4 L = kernel_path_integrate(kg, &rng, sample, ray, throughput)/throughput;
#else
float3 throughput = make_float3(1.0f, 1.0f, 1.0f);
float4 L = kernel_path_integrate(kg, &rng, sample, ray, throughput);
#endif
/* accumulate result in output buffer */ /* accumulate result in output buffer */
int index = offset + x + y*stride; kernel_write_pass_float4(buffer, sample, L);
if(sample == 0) path_rng_end(kg, rng_state, rng);
buffer[index] = L;
else
buffer[index] += L;
path_rng_end(kg, rng_state, rng, x, y, offset, stride);
} }
CCL_NAMESPACE_END CCL_NAMESPACE_END

44
intern/cycles/kernel/kernel_random.h
View File

@@ -123,7 +123,7 @@ __device_inline float path_rng(KernelGlobals *kg, RNG *rng, int sample, int dime
#endif #endif
} }
__device_inline void path_rng_init(KernelGlobals *kg, __global uint *rng_state, int sample, RNG *rng, int x, int y, int offset, int stride, float *fx, float *fy) __device_inline void path_rng_init(KernelGlobals *kg, __global uint *rng_state, int sample, RNG *rng, int x, int y, float *fx, float *fy)
{ {
#ifdef __SOBOL_FULL_SCREEN__ #ifdef __SOBOL_FULL_SCREEN__
uint px, py; uint px, py;
@@ -135,19 +135,31 @@ __device_inline void path_rng_init(KernelGlobals *kg, __global uint *rng_state,
*rng ^= kernel_data.integrator.seed; *rng ^= kernel_data.integrator.seed;
*fx = size * (float)px * (1.0f/(float)0xFFFFFFFF) - x; if(sample == 0) {
*fy = size * (float)py * (1.0f/(float)0xFFFFFFFF) - y; *fx = 0.5f;
*fy = 0.5f;
}
else {
*fx = size * (float)px * (1.0f/(float)0xFFFFFFFF) - x;
*fy = size * (float)py * (1.0f/(float)0xFFFFFFFF) - y;
}
#else #else
*rng = rng_state[offset + x + y*stride]; *rng = *rng_state;
*rng ^= kernel_data.integrator.seed; *rng ^= kernel_data.integrator.seed;
*fx = path_rng(kg, rng, sample, PRNG_FILTER_U); if(sample == 0) {
*fy = path_rng(kg, rng, sample, PRNG_FILTER_V); *fx = 0.5f;
*fy = 0.5f;
}
else {
*fx = path_rng(kg, rng, sample, PRNG_FILTER_U);
*fy = path_rng(kg, rng, sample, PRNG_FILTER_V);
}
#endif #endif
} }
__device void path_rng_end(KernelGlobals *kg, __global uint *rng_state, RNG rng, int x, int y, int offset, int stride) __device void path_rng_end(KernelGlobals *kg, __global uint *rng_state, RNG rng)
{ {
/* nothing to do */ /* nothing to do */
} }
@@ -163,21 +175,27 @@ __device float path_rng(KernelGlobals *kg, RNG *rng, int sample, int dimension)
return (float)*rng * (1.0f/(float)0xFFFFFFFF); return (float)*rng * (1.0f/(float)0xFFFFFFFF);
} }
__device void path_rng_init(KernelGlobals *kg, __global uint *rng_state, int sample, RNG *rng, int x, int y, int offset, int stride, float *fx, float *fy) __device void path_rng_init(KernelGlobals *kg, __global uint *rng_state, int sample, RNG *rng, int x, int y, float *fx, float *fy)
{ {
/* load state */ /* load state */
*rng = rng_state[offset + x + y*stride]; *rng = *rng_state;
*rng ^= kernel_data.integrator.seed; *rng ^= kernel_data.integrator.seed;
*fx = path_rng(kg, rng, sample, PRNG_FILTER_U); if(sample == 0) {
*fy = path_rng(kg, rng, sample, PRNG_FILTER_V); *fx = 0.5f;
*fy = 0.5f;
}
else {
*fx = path_rng(kg, rng, sample, PRNG_FILTER_U);
*fy = path_rng(kg, rng, sample, PRNG_FILTER_V);
}
} }
__device void path_rng_end(KernelGlobals *kg, __global uint *rng_state, RNG rng, int x, int y, int offset, int stride) __device void path_rng_end(KernelGlobals *kg, __global uint *rng_state, RNG rng)
{ {
/* store state for next sample */ /* store state for next sample */
rng_state[offset + x + y*stride] = rng; *rng_state = rng;
} }
#endif #endif

108
intern/cycles/kernel/kernel_shader.h
View File

@@ -75,7 +75,7 @@ __device_inline void shader_setup_from_ray(KernelGlobals *kg, ShaderData *sd,
if(sd->shader & SHADER_SMOOTH_NORMAL) if(sd->shader & SHADER_SMOOTH_NORMAL)
sd->N = triangle_smooth_normal(kg, sd->prim, sd->u, sd->v); sd->N = triangle_smooth_normal(kg, sd->prim, sd->u, sd->v);
sd->flag = kernel_tex_fetch(__shader_flag, sd->shader & SHADER_MASK); sd->flag = kernel_tex_fetch(__shader_flag, (sd->shader & SHADER_MASK)*2);
#ifdef __DPDU__ #ifdef __DPDU__
/* dPdu/dPdv */ /* dPdu/dPdv */
@@ -166,7 +166,7 @@ __device void shader_setup_from_sample(KernelGlobals *kg, ShaderData *sd,
#endif #endif
} }
sd->flag = kernel_tex_fetch(__shader_flag, sd->shader & SHADER_MASK); sd->flag = kernel_tex_fetch(__shader_flag, (sd->shader & SHADER_MASK)*2);
#ifdef __DPDU__ #ifdef __DPDU__
/* dPdu/dPdv */ /* dPdu/dPdv */
@@ -243,7 +243,7 @@ __device_inline void shader_setup_from_background(KernelGlobals *kg, ShaderData
sd->Ng = -sd->P; sd->Ng = -sd->P;
sd->I = -sd->P; sd->I = -sd->P;
sd->shader = kernel_data.background.shader; sd->shader = kernel_data.background.shader;
sd->flag = kernel_tex_fetch(__shader_flag, sd->shader & SHADER_MASK); sd->flag = kernel_tex_fetch(__shader_flag, (sd->shader & SHADER_MASK)*2);
#ifdef __INSTANCING__ #ifdef __INSTANCING__
sd->object = ~0; sd->object = ~0;
@@ -275,8 +275,8 @@ __device_inline void shader_setup_from_background(KernelGlobals *kg, ShaderData
#ifdef __MULTI_CLOSURE__ #ifdef __MULTI_CLOSURE__
__device_inline float3 _shader_bsdf_multi_eval(const ShaderData *sd, const float3 omega_in, float *pdf, __device_inline void _shader_bsdf_multi_eval(const ShaderData *sd, const float3 omega_in, float *pdf,
int skip_bsdf, float3 sum_eval, float sum_pdf, float sum_sample_weight) int skip_bsdf, BsdfEval *bsdf_eval, float sum_pdf, float sum_sample_weight)
{ {
for(int i = 0; i< sd->num_closure; i++) { for(int i = 0; i< sd->num_closure; i++) {
if(i == skip_bsdf) if(i == skip_bsdf)
@@ -293,7 +293,7 @@ __device_inline float3 _shader_bsdf_multi_eval(const ShaderData *sd, const float
#endif #endif
if(bsdf_pdf != 0.0f) { if(bsdf_pdf != 0.0f) {
sum_eval += eval*sc->weight; bsdf_eval_accum(bsdf_eval, sc->type, eval*sc->weight);
sum_pdf += bsdf_pdf*sc->sample_weight; sum_pdf += bsdf_pdf*sc->sample_weight;
} }
@@ -302,25 +302,27 @@ __device_inline float3 _shader_bsdf_multi_eval(const ShaderData *sd, const float
} }
*pdf = sum_pdf/sum_sample_weight; *pdf = sum_pdf/sum_sample_weight;
return sum_eval;
} }
#endif #endif
__device float3 shader_bsdf_eval(KernelGlobals *kg, const ShaderData *sd, __device void shader_bsdf_eval(KernelGlobals *kg, const ShaderData *sd,
const float3 omega_in, float *pdf) const float3 omega_in, BsdfEval *eval, float *pdf)
{ {
#ifdef __MULTI_CLOSURE__ #ifdef __MULTI_CLOSURE__
return _shader_bsdf_multi_eval(sd, omega_in, pdf, -1, make_float3(0.0f, 0.0f, 0.0f), 0.0f, 0.0f); bsdf_eval_init(eval, NBUILTIN_CLOSURES, make_float3(0.0f, 0.0f, 0.0f), kernel_data.film.use_light_pass);
return _shader_bsdf_multi_eval(sd, omega_in, pdf, -1, eval, 0.0f, 0.0f);
#else #else
const ShaderClosure *sc = &sd->closure; const ShaderClosure *sc = &sd->closure;
*pdf = 0.0f; *pdf = 0.0f;
return svm_bsdf_eval(sd, sc, omega_in, pdf)*sc->weight; *eval = svm_bsdf_eval(sd, sc, omega_in, pdf)*sc->weight;
#endif #endif
} }
__device int shader_bsdf_sample(KernelGlobals *kg, const ShaderData *sd, __device int shader_bsdf_sample(KernelGlobals *kg, const ShaderData *sd,
float randu, float randv, float3 *eval, float randu, float randv, BsdfEval *bsdf_eval,
float3 *omega_in, differential3 *domega_in, float *pdf) float3 *omega_in, differential3 *domega_in, float *pdf)
{ {
#ifdef __MULTI_CLOSURE__ #ifdef __MULTI_CLOSURE__
@@ -359,27 +361,28 @@ __device int shader_bsdf_sample(KernelGlobals *kg, const ShaderData *sd,
const ShaderClosure *sc = &sd->closure[sampled]; const ShaderClosure *sc = &sd->closure[sampled];
int label; int label;
float3 eval;
*pdf = 0.0f; *pdf = 0.0f;
#ifdef __OSL__ #ifdef __OSL__
label = OSLShader::bsdf_sample(sd, sc, randu, randv, *eval, *omega_in, *domega_in, *pdf); label = OSLShader::bsdf_sample(sd, sc, randu, randv, eval, *omega_in, *domega_in, *pdf);
#else #else
label = svm_bsdf_sample(sd, sc, randu, randv, eval, omega_in, domega_in, pdf); label = svm_bsdf_sample(sd, sc, randu, randv, &eval, omega_in, domega_in, pdf);
#endif #endif
*eval *= sc->weight; bsdf_eval_init(bsdf_eval, sc->type, eval*sc->weight, kernel_data.film.use_light_pass);
if(sd->num_closure > 1 && *pdf != 0.0f) { if(sd->num_closure > 1 && *pdf != 0.0f) {
float sweight = sc->sample_weight; float sweight = sc->sample_weight;
*eval = _shader_bsdf_multi_eval(sd, *omega_in, pdf, sampled, *eval, *pdf*sweight, sweight); _shader_bsdf_multi_eval(sd, *omega_in, pdf, sampled, bsdf_eval, *pdf*sweight, sweight);
} }
return label; return label;
#else #else
/* sample the single closure that we picked */ /* sample the single closure that we picked */
*pdf = 0.0f; *pdf = 0.0f;
int label = svm_bsdf_sample(sd, &sd->closure, randu, randv, eval, omega_in, domega_in, pdf); int label = svm_bsdf_sample(sd, &sd->closure, randu, randv, bsdf_eval, omega_in, domega_in, pdf);
*eval *= sd->closure.weight; *bsdf_eval *= sd->closure.weight;
return label; return label;
#endif #endif
} }
@@ -421,6 +424,68 @@ __device float3 shader_bsdf_transparency(KernelGlobals *kg, ShaderData *sd)
#endif #endif
} }
__device float3 shader_bsdf_diffuse(KernelGlobals *kg, ShaderData *sd)
{
#ifdef __MULTI_CLOSURE__
float3 eval = make_float3(0.0f, 0.0f, 0.0f);
for(int i = 0; i< sd->num_closure; i++) {
ShaderClosure *sc = &sd->closure[i];
if(CLOSURE_IS_BSDF_DIFFUSE(sc->type))
eval += sc->weight;
}
return eval;
#else
if(CLOSURE_IS_BSDF_DIFFUSE(sd->closure.type))
return sd->closure.weight;
else
return make_float3(0.0f, 0.0f, 0.0f);
#endif
}
__device float3 shader_bsdf_glossy(KernelGlobals *kg, ShaderData *sd)
{
#ifdef __MULTI_CLOSURE__
float3 eval = make_float3(0.0f, 0.0f, 0.0f);
for(int i = 0; i< sd->num_closure; i++) {
ShaderClosure *sc = &sd->closure[i];
if(CLOSURE_IS_BSDF_GLOSSY(sc->type))
eval += sc->weight;
}
return eval;
#else
if(CLOSURE_IS_BSDF_GLOSSY(sd->closure.type))
return sd->closure.weight;
else
return make_float3(0.0f, 0.0f, 0.0f);
#endif
}
__device float3 shader_bsdf_transmission(KernelGlobals *kg, ShaderData *sd)
{
#ifdef __MULTI_CLOSURE__
float3 eval = make_float3(0.0f, 0.0f, 0.0f);
for(int i = 0; i< sd->num_closure; i++) {
ShaderClosure *sc = &sd->closure[i];
if(CLOSURE_IS_BSDF_TRANSMISSION(sc->type))
eval += sc->weight;
}
return eval;
#else
if(CLOSURE_IS_BSDF_TRANSMISSION(sd->closure.type))
return sd->closure.weight;
else
return make_float3(0.0f, 0.0f, 0.0f);
#endif
}
/* Emission */ /* Emission */
@@ -588,12 +653,17 @@ __device bool shader_transparent_shadow(KernelGlobals *kg, Intersection *isect)
int prim = kernel_tex_fetch(__prim_index, isect->prim); int prim = kernel_tex_fetch(__prim_index, isect->prim);
float4 Ns = kernel_tex_fetch(__tri_normal, prim); float4 Ns = kernel_tex_fetch(__tri_normal, prim);
int shader = __float_as_int(Ns.w); int shader = __float_as_int(Ns.w);
int flag = kernel_tex_fetch(__shader_flag, shader & SHADER_MASK); int flag = kernel_tex_fetch(__shader_flag, (shader & SHADER_MASK)*2);
return (flag & SD_HAS_SURFACE_TRANSPARENT) != 0; return (flag & SD_HAS_SURFACE_TRANSPARENT) != 0;
} }
#endif #endif
__device int shader_pass_id(KernelGlobals *kg, ShaderData *sd)
{
return kernel_tex_fetch(__shader_flag, (sd->shader & SHADER_MASK)*2 + 1);
}
/* Free ShaderData */ /* Free ShaderData */
__device void shader_release(KernelGlobals *kg, ShaderData *sd) __device void shader_release(KernelGlobals *kg, ShaderData *sd)

99
intern/cycles/kernel/kernel_types.h
View File

@@ -70,6 +70,9 @@ CCL_NAMESPACE_BEGIN
#ifdef __KERNEL_ADV_SHADING__ #ifdef __KERNEL_ADV_SHADING__
#define __MULTI_CLOSURE__ #define __MULTI_CLOSURE__
#define __TRANSPARENT_SHADOWS__ #define __TRANSPARENT_SHADOWS__
#ifdef __KERNEL_CPU__
#define __PASSES__
#endif
#endif #endif
//#define __MULTI_LIGHT__ //#define __MULTI_LIGHT__
@@ -150,6 +153,75 @@ typedef enum ClosureLabel {
LABEL_STOP = 2048 LABEL_STOP = 2048
} ClosureLabel; } ClosureLabel;
/* Render Passes */
typedef enum PassType {
PASS_NONE = 0,
PASS_COMBINED = 1,
PASS_DEPTH = 2,
PASS_NORMAL = 8,
PASS_UV = 16,
PASS_OBJECT_ID = 32,
PASS_MATERIAL_ID = 64,
PASS_DIFFUSE_COLOR = 128,
PASS_GLOSSY_COLOR = 256,
PASS_TRANSMISSION_COLOR = 512,
PASS_DIFFUSE_INDIRECT = 1024,
PASS_GLOSSY_INDIRECT = 2048,
PASS_TRANSMISSION_INDIRECT = 4096,
PASS_DIFFUSE_DIRECT = 8192,
PASS_GLOSSY_DIRECT = 16384,
PASS_TRANSMISSION_DIRECT = 32768,
PASS_EMISSION = 65536,
PASS_BACKGROUND = 131072
} PassType;
#define PASS_ALL (~0)
#ifdef __PASSES__
typedef float3 PathThroughput;
struct PathRadiance {
int use_light_pass;
float3 indirect;
float3 direct_throughput;
float3 direct_emission;
float3 color_diffuse;
float3 color_glossy;
float3 color_transmission;
float3 direct_diffuse;
float3 direct_glossy;
float3 direct_transmission;
float3 indirect_diffuse;
float3 indirect_glossy;
float3 indirect_transmission;
float3 emission;
float3 background;
};
struct BsdfEval {
int use_light_pass;
float3 diffuse;
float3 glossy;
float3 transmission;
float3 transparent;
};
#else
typedef float3 PathThroughput;
typedef float3 PathRadiance;
typedef float3 BsdfEval;
#endif
/* Shader Flag */ /* Shader Flag */
typedef enum ShaderFlag { typedef enum ShaderFlag {
@@ -353,7 +425,32 @@ typedef struct KernelCamera {
typedef struct KernelFilm { typedef struct KernelFilm {
float exposure; float exposure;
int pad1, pad2, pad3; int pass_flag;
int pass_stride;
int use_light_pass;
int pass_combined;
int pass_depth;
int pass_normal;
int pass_pad;
int pass_uv;
int pass_object_id;
int pass_material_id;
int pass_diffuse_color;
int pass_glossy_color;
int pass_transmission_color;
int pass_diffuse_indirect;
int pass_glossy_indirect;
int pass_transmission_indirect;
int pass_diffuse_direct;
int pass_glossy_direct;
int pass_transmission_direct;
int pass_emission;
int pass_background;
} KernelFilm; } KernelFilm;
typedef struct KernelBackground { typedef struct KernelBackground {

25
intern/cycles/kernel/svm/svm_types.h
View File

@@ -266,22 +266,26 @@ typedef enum ShaderType {
typedef enum ClosureType { typedef enum ClosureType {
CLOSURE_BSDF_ID, CLOSURE_BSDF_ID,
CLOSURE_BSDF_DIFFUSE_ID, CLOSURE_BSDF_DIFFUSE_ID,
CLOSURE_BSDF_OREN_NAYAR_ID, CLOSURE_BSDF_OREN_NAYAR_ID,
CLOSURE_BSDF_TRANSLUCENT_ID,
CLOSURE_BSDF_REFLECTION_ID, CLOSURE_BSDF_REFLECTION_ID,
CLOSURE_BSDF_REFRACTION_ID,
CLOSURE_BSDF_GLASS_ID,
CLOSURE_BSDF_TRANSPARENT_ID,
CLOSURE_BSDF_MICROFACET_GGX_ID, CLOSURE_BSDF_MICROFACET_GGX_ID,
CLOSURE_BSDF_MICROFACET_GGX_REFRACTION_ID,
CLOSURE_BSDF_MICROFACET_BECKMANN_ID, CLOSURE_BSDF_MICROFACET_BECKMANN_ID,
CLOSURE_BSDF_MICROFACET_BECKMANN_REFRACTION_ID,
CLOSURE_BSDF_WARD_ID, CLOSURE_BSDF_WARD_ID,
CLOSURE_BSDF_ASHIKHMIN_VELVET_ID,
CLOSURE_BSDF_WESTIN_BACKSCATTER_ID,
CLOSURE_BSDF_WESTIN_SHEEN_ID, CLOSURE_BSDF_WESTIN_SHEEN_ID,
CLOSURE_BSDF_TRANSLUCENT_ID,
CLOSURE_BSDF_REFRACTION_ID,
CLOSURE_BSDF_WESTIN_BACKSCATTER_ID,
CLOSURE_BSDF_ASHIKHMIN_VELVET_ID,
CLOSURE_BSDF_MICROFACET_BECKMANN_REFRACTION_ID,
CLOSURE_BSDF_MICROFACET_GGX_REFRACTION_ID,
CLOSURE_BSDF_GLASS_ID,
CLOSURE_BSDF_TRANSPARENT_ID,
CLOSURE_BSSRDF_CUBIC_ID, CLOSURE_BSSRDF_CUBIC_ID,
CLOSURE_EMISSION_ID, CLOSURE_EMISSION_ID,
CLOSURE_DEBUG_ID, CLOSURE_DEBUG_ID,
@@ -297,7 +301,10 @@ typedef enum ClosureType {
} ClosureType; } ClosureType;
/* watch this, being lazy with memory usage */ /* watch this, being lazy with memory usage */
#define CLOSURE_IS_BSDF(type) (type <= CLOSURE_BSDF_WESTIN_SHEEN_ID) #define CLOSURE_IS_BSDF(type) (type <= CLOSURE_BSDF_TRANSPARENT_ID)
#define CLOSURE_IS_BSDF_DIFFUSE(type) (type >= CLOSURE_BSDF_DIFFUSE_ID && type <= CLOSURE_BSDF_OREN_NAYAR_ID)
#define CLOSURE_IS_BSDF_GLOSSY(type) (type >= CLOSURE_BSDF_REFLECTION_ID && type <= CLOSURE_BSDF_WESTIN_SHEEN_ID)
#define CLOSURE_IS_BSDF_TRANSMISSION(type) (type >= CLOSURE_BSDF_TRANSLUCENT_ID && type <= CLOSURE_BSDF_GLASS_ID)
#define CLOSURE_IS_VOLUME(type) (type >= CLOSURE_VOLUME_ID && type <= CLOSURE_VOLUME_ISOTROPIC_ID) #define CLOSURE_IS_VOLUME(type) (type >= CLOSURE_VOLUME_ID && type <= CLOSURE_VOLUME_ISOTROPIC_ID)
#define CLOSURE_IS_EMISSION(type) (type == CLOSURE_EMISSION_ID) #define CLOSURE_IS_EMISSION(type) (type == CLOSURE_EMISSION_ID)
#define CLOSURE_IS_HOLDOUT(type) (type == CLOSURE_HOLDOUT_ID) #define CLOSURE_IS_HOLDOUT(type) (type == CLOSURE_HOLDOUT_ID)

120
intern/cycles/render/buffers.cpp
View File

@@ -22,6 +22,7 @@
#include "device.h" #include "device.h"
#include "util_debug.h" #include "util_debug.h"
#include "util_foreach.h"
#include "util_hash.h" #include "util_hash.h"
#include "util_image.h" #include "util_image.h"
#include "util_math.h" #include "util_math.h"
@@ -31,6 +32,48 @@
CCL_NAMESPACE_BEGIN CCL_NAMESPACE_BEGIN
/* Buffer Params */
BufferParams::BufferParams()
{
width = 0;
height = 0;
full_x = 0;
full_y = 0;
full_width = 0;
full_height = 0;
Pass::add(PASS_COMBINED, passes);
}
void BufferParams::get_offset_stride(int& offset, int& stride)
{
offset = -(full_x + full_y*width);
stride = width;
}
bool BufferParams::modified(const BufferParams& params)
{
return !(full_x == params.full_x
&& full_y == params.full_y
&& width == params.width
&& height == params.height
&& full_width == params.full_width
&& full_height == params.full_height
&& Pass::equals(passes, params.passes));
}
int BufferParams::get_passes_size()
{
int size = 0;
foreach(Pass& pass, passes)
size += pass.components;
return size;
}
/* Render Buffers */ /* Render Buffers */
RenderBuffers::RenderBuffers(Device *device_) RenderBuffers::RenderBuffers(Device *device_)
@@ -64,7 +107,7 @@ void RenderBuffers::reset(Device *device, BufferParams& params_)
device_free(); device_free();
/* allocate buffer */ /* allocate buffer */
buffer.resize(params.width, params.height); buffer.resize(params.width*params.height*params.get_passes_size());
device->mem_alloc(buffer, MEM_READ_WRITE); device->mem_alloc(buffer, MEM_READ_WRITE);
device->mem_zero(buffer); device->mem_zero(buffer);
@@ -82,31 +125,76 @@ void RenderBuffers::reset(Device *device, BufferParams& params_)
device->mem_copy_to(rng_state); device->mem_copy_to(rng_state);
} }
float4 *RenderBuffers::copy_from_device(float exposure, int sample) bool RenderBuffers::copy_from_device()
{ {
if(!buffer.device_pointer) if(!buffer.device_pointer)
return NULL; return false;
device->mem_copy_from(buffer, 0, params.width, params.height, sizeof(float4)); device->mem_copy_from(buffer, 0, params.width, params.height, sizeof(float4));
float4 *out = new float4[params.width*params.height]; return true;
float4 *in = (float4*)buffer.data_pointer; }
float scale = 1.0f/(float)sample;
for(int i = params.width*params.height - 1; i >= 0; i--) {
float4 rgba = in[i]*scale;
rgba.x = rgba.x*exposure; bool RenderBuffers::get_pass(PassType type, float exposure, int sample, int components, float *pixels)
rgba.y = rgba.y*exposure; {
rgba.z = rgba.z*exposure; int pass_offset = 0;
/* clamp since alpha might be > 1.0 due to russian roulette */ foreach(Pass& pass, params.passes) {
rgba.w = clamp(rgba.w, 0.0f, 1.0f); if(pass.type != type) {
pass_offset += pass.components;
continue;
}
out[i] = rgba; float *in = (float*)buffer.data_pointer + pass_offset;
int pass_stride = params.get_passes_size();
float scale = (pass.filter)? 1.0f/(float)sample: 1.0f;
float scale_exposure = (pass.exposure)? scale*exposure: scale;
int size = params.width*params.height;
if(components == 1) {
assert(pass.components == components);
/* scalar */
for(int i = 0; i < size; i++, in += pass_stride, pixels++) {
float f = *in;
pixels[0] = f*scale_exposure;
}
}
else if(components == 3) {
assert(pass.components == 4);
/* RGB/vector */
for(int i = 0; i < size; i++, in += pass_stride, pixels += 3) {
float3 f = make_float3(in[0], in[1], in[2]);
pixels[0] = f.x*scale_exposure;
pixels[1] = f.y*scale_exposure;
pixels[2] = f.z*scale_exposure;
}
}
else if(components == 4) {
assert(pass.components == components);
/* RGBA */
for(int i = 0; i < size; i++, in += pass_stride, pixels += 4) {
float4 f = make_float4(in[0], in[1], in[2], in[3]);
pixels[0] = f.x*scale_exposure;
pixels[1] = f.y*scale_exposure;
pixels[2] = f.z*scale_exposure;
/* clamp since alpha might be > 1.0 due to russian roulette */
pixels[3] = clamp(f.w*scale, 0.0f, 1.0f);
}
}
return true;
} }
return out; return false;
} }
/* Display Buffer */ /* Display Buffer */

42
intern/cycles/render/buffers.h
View File

@@ -21,6 +21,10 @@
#include "device_memory.h" #include "device_memory.h"
#include "film.h"
#include "kernel_types.h"
#include "util_string.h" #include "util_string.h"
#include "util_thread.h" #include "util_thread.h"
#include "util_types.h" #include "util_types.h"
@@ -45,32 +49,16 @@ public:
int full_width; int full_width;
int full_height; int full_height;
BufferParams() /* passes */
{ vector<Pass> passes;
width = 0;
height = 0;
full_x = 0; /* functions */
full_y = 0; BufferParams();
full_width = 0;
full_height = 0;
}
void get_offset_stride(int& offset, int& stride) void get_offset_stride(int& offset, int& stride);
{ bool modified(const BufferParams& params);
offset = -(full_x + full_y*width); void add_pass(PassType type);
stride = width; int get_passes_size();
}
bool modified(const BufferParams& params)
{
return !(full_x == params.full_x
&& full_y == params.full_y
&& width == params.width
&& height == params.height
&& full_width == params.full_width
&& full_height == params.full_height);
}
}; };
/* Render Buffers */ /* Render Buffers */
@@ -80,7 +68,7 @@ public:
/* buffer parameters */ /* buffer parameters */
BufferParams params; BufferParams params;
/* float buffer */ /* float buffer */
device_vector<float4> buffer; device_vector<float> buffer;
/* random number generator state */ /* random number generator state */
device_vector<uint> rng_state; device_vector<uint> rng_state;
/* mutex, must be locked manually by callers */ /* mutex, must be locked manually by callers */
@@ -90,7 +78,9 @@ public:
~RenderBuffers(); ~RenderBuffers();
void reset(Device *device, BufferParams& params); void reset(Device *device, BufferParams& params);
float4 *copy_from_device(float exposure, int sample);
bool copy_from_device();
bool get_pass(PassType type, float exposure, int sample, int components, float *pixels);
protected: protected:
void device_free(); void device_free();

179
intern/cycles/render/film.cpp
View File

@@ -21,11 +21,111 @@
#include "film.h" #include "film.h"
#include "scene.h" #include "scene.h"
#include "util_foreach.h"
CCL_NAMESPACE_BEGIN CCL_NAMESPACE_BEGIN
/* Pass */
void Pass::add(PassType type, vector<Pass>& passes)
{
Pass pass;
pass.type = type;
pass.filter = true;
pass.exposure = false;
switch(type) {
case PASS_NONE:
pass.components = 0;
break;
case PASS_COMBINED:
pass.components = 4;
pass.exposure = true;
break;
case PASS_DEPTH:
pass.components = 1;
pass.filter = false;
break;
case PASS_NORMAL:
pass.components = 4;
break;
case PASS_UV:
pass.components = 4;
break;
case PASS_OBJECT_ID:
pass.components = 1;
pass.filter = false;
break;
case PASS_MATERIAL_ID:
pass.components = 1;
pass.filter = false;
break;
case PASS_DIFFUSE_COLOR:
pass.components = 4;
break;
case PASS_GLOSSY_COLOR:
pass.components = 4;
break;
case PASS_TRANSMISSION_COLOR:
pass.components = 4;
break;
case PASS_DIFFUSE_INDIRECT:
pass.components = 4;
pass.exposure = true;
break;
case PASS_GLOSSY_INDIRECT:
pass.components = 4;
pass.exposure = true;
break;
case PASS_TRANSMISSION_INDIRECT:
pass.components = 4;
pass.exposure = true;
break;
case PASS_DIFFUSE_DIRECT:
pass.components = 4;
pass.exposure = true;
break;
case PASS_GLOSSY_DIRECT:
pass.components = 4;
pass.exposure = true;
break;
case PASS_TRANSMISSION_DIRECT:
pass.components = 4;
pass.exposure = true;
break;
case PASS_EMISSION:
pass.components = 4;
pass.exposure = true;
break;
case PASS_BACKGROUND:
pass.components = 4;
pass.exposure = true;
break;
}
passes.push_back(pass);
}
bool Pass::equals(const vector<Pass>& A, const vector<Pass>& B)
{
if(A.size() != B.size())
return false;
for(int i = 0; i < A.size(); i++)
if(A[i].type != B[i].type)
return false;
return true;
}
/* Film */
Film::Film() Film::Film()
{ {
exposure = 0.8f; exposure = 0.8f;
Pass::add(PASS_COMBINED, passes);
need_update = true; need_update = true;
} }
@@ -42,6 +142,82 @@ void Film::device_update(Device *device, DeviceScene *dscene)
/* update __data */ /* update __data */
kfilm->exposure = exposure; kfilm->exposure = exposure;
kfilm->pass_flag = 0;
kfilm->pass_stride = 0;
kfilm->use_light_pass = 0;
foreach(Pass& pass, passes) {
kfilm->pass_flag |= pass.type;
switch(pass.type) {
case PASS_COMBINED:
kfilm->pass_combined = kfilm->pass_stride;
break;
case PASS_DEPTH:
kfilm->pass_depth = kfilm->pass_stride;
break;
case PASS_NORMAL:
kfilm->pass_normal = kfilm->pass_stride;
break;
case PASS_UV:
kfilm->pass_uv = kfilm->pass_stride;
break;
case PASS_OBJECT_ID:
kfilm->pass_object_id = kfilm->pass_stride;
break;
case PASS_MATERIAL_ID:
kfilm->pass_material_id = kfilm->pass_stride;
break;
case PASS_DIFFUSE_COLOR:
kfilm->pass_diffuse_color = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_GLOSSY_COLOR:
kfilm->pass_glossy_color = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_TRANSMISSION_COLOR:
kfilm->pass_transmission_color = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_DIFFUSE_INDIRECT:
kfilm->pass_diffuse_indirect = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_GLOSSY_INDIRECT:
kfilm->pass_glossy_indirect = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_TRANSMISSION_INDIRECT:
kfilm->pass_transmission_indirect = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_DIFFUSE_DIRECT:
kfilm->pass_diffuse_direct = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_GLOSSY_DIRECT:
kfilm->pass_glossy_direct = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_TRANSMISSION_DIRECT:
kfilm->pass_transmission_direct = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_EMISSION:
kfilm->pass_emission = kfilm->pass_stride;
kfilm->use_light_pass = 1;
break;
case PASS_BACKGROUND:
kfilm->pass_background = kfilm->pass_stride;
kfilm->use_light_pass = 1;
case PASS_NONE:
break;
}
kfilm->pass_stride += pass.components;
}
need_update = false; need_update = false;
} }
@@ -52,7 +228,8 @@ void Film::device_free(Device *device, DeviceScene *dscene)
bool Film::modified(const Film& film) bool Film::modified(const Film& film)
{ {
return !(exposure == film.exposure); return !(exposure == film.exposure
&& Pass::equals(passes, film.passes));
} }
void Film::tag_update(Scene *scene) void Film::tag_update(Scene *scene)

15
intern/cycles/render/film.h
View File

@@ -20,6 +20,9 @@
#define __FILM_H__ #define __FILM_H__
#include "util_string.h" #include "util_string.h"
#include "util_vector.h"
#include "kernel_types.h"
CCL_NAMESPACE_BEGIN CCL_NAMESPACE_BEGIN
@@ -27,9 +30,21 @@ class Device;
class DeviceScene; class DeviceScene;
class Scene; class Scene;
class Pass {
public:
PassType type;
int components;
bool filter;
bool exposure;
static void add(PassType type, vector<Pass>& passes);
static bool equals(const vector<Pass>& A, const vector<Pass>& B);
};
class Film { class Film {
public: public:
float exposure; float exposure;
vector<Pass> passes;
bool need_update; bool need_update;
Film(); Film();

4
intern/cycles/render/object.cpp
View File

@@ -36,6 +36,7 @@ Object::Object()
mesh = NULL; mesh = NULL;
tfm = transform_identity(); tfm = transform_identity();
visibility = ~0; visibility = ~0;
pass_id = 0;
} }
Object::~Object() Object::~Object()
@@ -135,6 +136,7 @@ void ObjectManager::device_update_transforms(Device *device, DeviceScene *dscene
/* todo: correct for displacement, and move to a better place */ /* todo: correct for displacement, and move to a better place */
float uniform_scale; float uniform_scale;
float surface_area = 0.0f; float surface_area = 0.0f;
float pass_id = ob->pass_id;
if(transform_uniform_scale(tfm, uniform_scale)) { if(transform_uniform_scale(tfm, uniform_scale)) {
map<Mesh*, float>::iterator it = surface_area_map.find(mesh); map<Mesh*, float>::iterator it = surface_area_map.find(mesh);
@@ -171,7 +173,7 @@ void ObjectManager::device_update_transforms(Device *device, DeviceScene *dscene
memcpy(&objects[offset], &tfm, sizeof(float4)*4); memcpy(&objects[offset], &tfm, sizeof(float4)*4);
memcpy(&objects[offset+4], &itfm, sizeof(float4)*4); memcpy(&objects[offset+4], &itfm, sizeof(float4)*4);
memcpy(&objects[offset+8], &ntfm, sizeof(float4)*4); memcpy(&objects[offset+8], &ntfm, sizeof(float4)*4);
objects[offset+12] = make_float4(surface_area, 0.0f, 0.0f, 0.0f); objects[offset+12] = make_float4(surface_area, pass_id, 0.0f, 0.0f);
i++; i++;

1
intern/cycles/render/object.h
View File

@@ -41,6 +41,7 @@ public:
Transform tfm; Transform tfm;
BoundBox bounds; BoundBox bounds;
ustring name; ustring name;
int pass_id;
vector<ParamValue> attributes; vector<ParamValue> attributes;
uint visibility; uint visibility;

5
intern/cycles/render/shader.cpp
View File

@@ -35,6 +35,7 @@ CCL_NAMESPACE_BEGIN
Shader::Shader() Shader::Shader()
{ {
name = ""; name = "";
pass_id = 0;
graph = NULL; graph = NULL;
graph_bump = NULL; graph_bump = NULL;
@@ -167,7 +168,7 @@ void ShaderManager::device_update_common(Device *device, DeviceScene *dscene, Sc
if(scene->shaders.size() == 0) if(scene->shaders.size() == 0)
return; return;
uint shader_flag_size = scene->shaders.size()*2; uint shader_flag_size = scene->shaders.size()*4;
uint *shader_flag = dscene->shader_flag.resize(shader_flag_size); uint *shader_flag = dscene->shader_flag.resize(shader_flag_size);
uint i = 0; uint i = 0;
@@ -184,7 +185,9 @@ void ShaderManager::device_update_common(Device *device, DeviceScene *dscene, Sc
flag |= SD_HOMOGENEOUS_VOLUME; flag |= SD_HOMOGENEOUS_VOLUME;
shader_flag[i++] = flag; shader_flag[i++] = flag;
shader_flag[i++] = shader->pass_id;
shader_flag[i++] = flag; shader_flag[i++] = flag;
shader_flag[i++] = shader->pass_id;
} }
device->tex_alloc("__shader_flag", dscene->shader_flag); device->tex_alloc("__shader_flag", dscene->shader_flag);

1
intern/cycles/render/shader.h
View File

@@ -47,6 +47,7 @@ class Shader {
public: public:
/* name */ /* name */
string name; string name;
int pass_id;
/* shader graph */ /* shader graph */
ShaderGraph *graph; ShaderGraph *graph;

47
source/blender/blenkernel/BKE_node.h
View File

@@ -557,25 +557,34 @@ void ntreeGPUMaterialNodes(struct bNodeTree *ntree, struct GPUMaterial *mat);
/* ************** COMPOSITE NODES *************** */ /* ************** COMPOSITE NODES *************** */
/* output socket defines */ /* output socket defines */
#define RRES_OUT_IMAGE 0 #define RRES_OUT_IMAGE 0
#define RRES_OUT_ALPHA 1 #define RRES_OUT_ALPHA 1
#define RRES_OUT_Z 2 #define RRES_OUT_Z 2
#define RRES_OUT_NORMAL 3 #define RRES_OUT_NORMAL 3
#define RRES_OUT_UV 4 #define RRES_OUT_UV 4
#define RRES_OUT_VEC 5 #define RRES_OUT_VEC 5
#define RRES_OUT_RGBA 6 #define RRES_OUT_RGBA 6
#define RRES_OUT_DIFF 7 #define RRES_OUT_DIFF 7
#define RRES_OUT_SPEC 8 #define RRES_OUT_SPEC 8
#define RRES_OUT_SHADOW 9 #define RRES_OUT_SHADOW 9
#define RRES_OUT_AO 10 #define RRES_OUT_AO 10
#define RRES_OUT_REFLECT 11 #define RRES_OUT_REFLECT 11
#define RRES_OUT_REFRACT 12 #define RRES_OUT_REFRACT 12
#define RRES_OUT_INDIRECT 13 #define RRES_OUT_INDIRECT 13
#define RRES_OUT_INDEXOB 14 #define RRES_OUT_INDEXOB 14
#define RRES_OUT_INDEXMA 15 #define RRES_OUT_INDEXMA 15
#define RRES_OUT_MIST 16 #define RRES_OUT_MIST 16
#define RRES_OUT_EMIT 17 #define RRES_OUT_EMIT 17
#define RRES_OUT_ENV 18 #define RRES_OUT_ENV 18
#define RRES_OUT_DIFF_DIRECT 19
#define RRES_OUT_DIFF_INDIRECT 20
#define RRES_OUT_DIFF_COLOR 21
#define RRES_OUT_GLOSSY_DIRECT 22
#define RRES_OUT_GLOSSY_INDIRECT 23
#define RRES_OUT_GLOSSY_COLOR 24
#define RRES_OUT_TRANSM_DIRECT 25
#define RRES_OUT_TRANSM_INDIRECT 26
#define RRES_OUT_TRANSM_COLOR 27
/* note: types are needed to restore callbacks, don't change values */ /* note: types are needed to restore callbacks, don't change values */
#define CMP_NODE_VIEWER 201 #define CMP_NODE_VIEWER 201

47
source/blender/makesdna/DNA_scene_types.h
View File

@@ -202,25 +202,34 @@ typedef struct SceneRenderLayer {
#define SCE_LAY_NEG_ZMASK 0x80000 #define SCE_LAY_NEG_ZMASK 0x80000
/* srl->passflag */ /* srl->passflag */
#define SCE_PASS_COMBINED (1<<0) #define SCE_PASS_COMBINED (1<<0)
#define SCE_PASS_Z (1<<1) #define SCE_PASS_Z (1<<1)
#define SCE_PASS_RGBA (1<<2) #define SCE_PASS_RGBA (1<<2)
#define SCE_PASS_DIFFUSE (1<<3) #define SCE_PASS_DIFFUSE (1<<3)
#define SCE_PASS_SPEC (1<<4) #define SCE_PASS_SPEC (1<<4)
#define SCE_PASS_SHADOW (1<<5) #define SCE_PASS_SHADOW (1<<5)
#define SCE_PASS_AO (1<<6) #define SCE_PASS_AO (1<<6)
#define SCE_PASS_REFLECT (1<<7) #define SCE_PASS_REFLECT (1<<7)
#define SCE_PASS_NORMAL (1<<8) #define SCE_PASS_NORMAL (1<<8)
#define SCE_PASS_VECTOR (1<<9) #define SCE_PASS_VECTOR (1<<9)
#define SCE_PASS_REFRACT (1<<10) #define SCE_PASS_REFRACT (1<<10)
#define SCE_PASS_INDEXOB (1<<11) #define SCE_PASS_INDEXOB (1<<11)
#define SCE_PASS_UV (1<<12) #define SCE_PASS_UV (1<<12)
#define SCE_PASS_INDIRECT (1<<13) #define SCE_PASS_INDIRECT (1<<13)
#define SCE_PASS_MIST (1<<14) #define SCE_PASS_MIST (1<<14)
#define SCE_PASS_RAYHITS (1<<15) #define SCE_PASS_RAYHITS (1<<15)
#define SCE_PASS_EMIT (1<<16) #define SCE_PASS_EMIT (1<<16)
#define SCE_PASS_ENVIRONMENT (1<<17) #define SCE_PASS_ENVIRONMENT (1<<17)
#define SCE_PASS_INDEXMA (1<<18) #define SCE_PASS_INDEXMA (1<<18)
#define SCE_PASS_DIFFUSE_DIRECT (1<<19)
#define SCE_PASS_DIFFUSE_INDIRECT (1<<20)
#define SCE_PASS_DIFFUSE_COLOR (1<<21)
#define SCE_PASS_GLOSSY_DIRECT (1<<22)
#define SCE_PASS_GLOSSY_INDIRECT (1<<23)
#define SCE_PASS_GLOSSY_COLOR (1<<24)
#define SCE_PASS_TRANSM_DIRECT (1<<25)
#define SCE_PASS_TRANSM_INDIRECT (1<<26)
#define SCE_PASS_TRANSM_COLOR (1<<27)
/* note, srl->passflag is treestore element 'nr' in outliner, short still... */ /* note, srl->passflag is treestore element 'nr' in outliner, short still... */

9
source/blender/makesrna/intern/rna_render.c
View File

@@ -472,6 +472,15 @@ static void rna_def_render_pass(BlenderRNA *brna)
{SCE_PASS_EMIT, "EMIT", 0, "Emit", ""}, {SCE_PASS_EMIT, "EMIT", 0, "Emit", ""},
{SCE_PASS_ENVIRONMENT, "ENVIRONMENT", 0, "Environment", ""}, {SCE_PASS_ENVIRONMENT, "ENVIRONMENT", 0, "Environment", ""},
{SCE_PASS_INDEXMA, "MATERIAL_INDEX", 0, "Material Index", ""}, {SCE_PASS_INDEXMA, "MATERIAL_INDEX", 0, "Material Index", ""},
{SCE_PASS_DIFFUSE_DIRECT, "DIFFUSE_DIRECT", 0, "Diffuse Direct", ""},
{SCE_PASS_DIFFUSE_INDIRECT, "DIFFUSE_INDIRECT", 0, "Diffuse Indirect", ""},
{SCE_PASS_DIFFUSE_COLOR, "DIFFUSE_COLOR", 0, "Diffuse Color", ""},
{SCE_PASS_GLOSSY_DIRECT, "GLOSSY_DIRECT", 0, "Glossy Direct", ""},
{SCE_PASS_GLOSSY_INDIRECT, "GLOSSY_INDIRECT", 0, "Glossy Indirect", ""},
{SCE_PASS_GLOSSY_COLOR, "GLOSSY_COLOR", 0, "Glossy Color", ""},
{SCE_PASS_TRANSM_DIRECT, "TRANSMISSION_DIRECT", 0, "Transmission Direct", ""},
{SCE_PASS_TRANSM_INDIRECT, "TRANSMISSION_INDIRECT", 0, "Transmission Indirect", ""},
{SCE_PASS_TRANSM_COLOR, "TRANSMISSION_COLOR", 0, "Transmission Color", ""},
{0, NULL, 0, NULL, NULL}}; {0, NULL, 0, NULL, NULL}};
srna= RNA_def_struct(brna, "RenderPass", NULL); srna= RNA_def_struct(brna, "RenderPass", NULL);

54
source/blender/makesrna/intern/rna_scene.c
View File

@@ -2065,6 +2065,60 @@ void rna_def_render_layer_common(StructRNA *srna, int scene)
RNA_def_property_ui_icon(prop, ICON_RESTRICT_RENDER_OFF, 1); RNA_def_property_ui_icon(prop, ICON_RESTRICT_RENDER_OFF, 1);
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update"); if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE); else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
prop= RNA_def_property(srna, "use_pass_diffuse_direct", PROP_BOOLEAN, PROP_NONE);
RNA_def_property_boolean_sdna(prop, NULL, "passflag", SCE_PASS_DIFFUSE_DIRECT);
RNA_def_property_ui_text(prop, "Diffuse Direct", "Deliver diffuse direct pass");
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
prop= RNA_def_property(srna, "use_pass_diffuse_indirect", PROP_BOOLEAN, PROP_NONE);
RNA_def_property_boolean_sdna(prop, NULL, "passflag", SCE_PASS_DIFFUSE_INDIRECT);
RNA_def_property_ui_text(prop, "Diffuse Indirect", "Deliver diffuse indirect pass");
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
prop= RNA_def_property(srna, "use_pass_diffuse_color", PROP_BOOLEAN, PROP_NONE);
RNA_def_property_boolean_sdna(prop, NULL, "passflag", SCE_PASS_DIFFUSE_COLOR);
RNA_def_property_ui_text(prop, "Diffuse Color", "Deliver diffuse color pass");
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
prop= RNA_def_property(srna, "use_pass_glossy_direct", PROP_BOOLEAN, PROP_NONE);
RNA_def_property_boolean_sdna(prop, NULL, "passflag", SCE_PASS_GLOSSY_DIRECT);
RNA_def_property_ui_text(prop, "Glossy Direct", "Deliver glossy direct pass");
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
prop= RNA_def_property(srna, "use_pass_glossy_indirect", PROP_BOOLEAN, PROP_NONE);
RNA_def_property_boolean_sdna(prop, NULL, "passflag", SCE_PASS_GLOSSY_INDIRECT);
RNA_def_property_ui_text(prop, "Glossy Indirect", "Deliver glossy indirect pass");
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
prop= RNA_def_property(srna, "use_pass_glossy_color", PROP_BOOLEAN, PROP_NONE);
RNA_def_property_boolean_sdna(prop, NULL, "passflag", SCE_PASS_GLOSSY_COLOR);
RNA_def_property_ui_text(prop, "Glossy Color", "Deliver glossy color pass");
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
prop= RNA_def_property(srna, "use_pass_transmission_direct", PROP_BOOLEAN, PROP_NONE);
RNA_def_property_boolean_sdna(prop, NULL, "passflag", SCE_PASS_TRANSM_DIRECT);
RNA_def_property_ui_text(prop, "Transmission Direct", "Deliver transmission direct pass");
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
prop= RNA_def_property(srna, "use_pass_transmission_indirect", PROP_BOOLEAN, PROP_NONE);
RNA_def_property_boolean_sdna(prop, NULL, "passflag", SCE_PASS_TRANSM_INDIRECT);
RNA_def_property_ui_text(prop, "Transmission Indirect", "Deliver transmission indirect pass");
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
prop= RNA_def_property(srna, "use_pass_transmission_color", PROP_BOOLEAN, PROP_NONE);
RNA_def_property_boolean_sdna(prop, NULL, "passflag", SCE_PASS_TRANSM_COLOR);
RNA_def_property_ui_text(prop, "Transmission Color", "Deliver transmission color pass");
if(scene) RNA_def_property_update(prop, NC_SCENE|ND_RENDER_OPTIONS, "rna_SceneRenderLayer_pass_update");
else RNA_def_property_clear_flag(prop, PROP_EDITABLE);
} }
static void rna_def_scene_game_recast_data(BlenderRNA *brna) static void rna_def_scene_game_recast_data(BlenderRNA *brna)

23
source/blender/nodes/composite/node_composite_tree.c
View File

@@ -694,7 +694,28 @@ static void force_hidden_passes(bNode *node, int passflag)
if(!(passflag & SCE_PASS_EMIT)) sock->flag |= SOCK_UNAVAIL; if(!(passflag & SCE_PASS_EMIT)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_ENV); sock= BLI_findlink(&node->outputs, RRES_OUT_ENV);
if(!(passflag & SCE_PASS_ENVIRONMENT)) sock->flag |= SOCK_UNAVAIL; if(!(passflag & SCE_PASS_ENVIRONMENT)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_DIFF_DIRECT);
if(!(passflag & SCE_PASS_DIFFUSE_DIRECT)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_DIFF_INDIRECT);
if(!(passflag & SCE_PASS_DIFFUSE_INDIRECT)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_DIFF_COLOR);
if(!(passflag & SCE_PASS_DIFFUSE_COLOR)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_GLOSSY_DIRECT);
if(!(passflag & SCE_PASS_GLOSSY_DIRECT)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_GLOSSY_INDIRECT);
if(!(passflag & SCE_PASS_GLOSSY_INDIRECT)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_GLOSSY_COLOR);
if(!(passflag & SCE_PASS_GLOSSY_COLOR)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_TRANSM_DIRECT);
if(!(passflag & SCE_PASS_TRANSM_DIRECT)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_TRANSM_INDIRECT);
if(!(passflag & SCE_PASS_TRANSM_INDIRECT)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_TRANSM_COLOR);
if(!(passflag & SCE_PASS_TRANSM_COLOR)) sock->flag |= SOCK_UNAVAIL;
sock= BLI_findlink(&node->outputs, RRES_OUT_TRANSM_COLOR);
} }
/* based on rules, force sockets hidden always */ /* based on rules, force sockets hidden always */

83
source/blender/nodes/composite/nodes/node_composite_image.c
View File

@@ -36,25 +36,34 @@
/* **************** IMAGE (and RenderResult, multilayer image) ******************** */ /* **************** IMAGE (and RenderResult, multilayer image) ******************** */
static bNodeSocketTemplate cmp_node_rlayers_out[]= { static bNodeSocketTemplate cmp_node_rlayers_out[]= {
{ SOCK_RGBA, 0, "Image", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Image", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_FLOAT, 0, "Alpha", 1.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_FLOAT, 0, "Alpha", 1.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_FLOAT, 0, "Z", 1.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_FLOAT, 0, "Z", 1.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_VECTOR, 0, "Normal", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_VECTOR, 0, "Normal", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_VECTOR, 0, "UV", 1.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_VECTOR, 0, "UV", 1.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_VECTOR, 0, "Speed", 1.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_VECTOR, 0, "Speed", 1.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Color", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Color", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Diffuse", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Diffuse", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Specular", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Specular", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Shadow", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Shadow", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "AO", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "AO", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Reflect", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Reflect", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Refract", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Refract", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Indirect", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Indirect", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_FLOAT, 0, "IndexOB", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_FLOAT, 0, "IndexOB", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_FLOAT, 0, "IndexMA", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_FLOAT, 0, "IndexMA", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_FLOAT, 0, "Mist", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_FLOAT, 0, "Mist", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Emit", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Emit", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Environment",0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f}, { SOCK_RGBA, 0, "Environment", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Diffuse Direct", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Diffuse Indirect", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Diffuse Color", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Glossy Direct", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Glossy Indirect", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Glossy Color", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Transmission Direct", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Transmission Indirect", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ SOCK_RGBA, 0, "Transmission Color", 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 1.0f},
{ -1, 0, "" } { -1, 0, "" }
}; };
@@ -238,6 +247,24 @@ static void outputs_multilayer_get(RenderData *rd, RenderLayer *rl, bNodeStack *
out[RRES_OUT_EMIT]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_EMIT); out[RRES_OUT_EMIT]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_EMIT);
if(out[RRES_OUT_ENV]->hasoutput) if(out[RRES_OUT_ENV]->hasoutput)
out[RRES_OUT_ENV]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_ENVIRONMENT); out[RRES_OUT_ENV]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_ENVIRONMENT);
if(out[RRES_OUT_DIFF_DIRECT]->hasoutput)
out[RRES_OUT_DIFF_DIRECT]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_DIFFUSE_DIRECT);
if(out[RRES_OUT_DIFF_INDIRECT]->hasoutput)
out[RRES_OUT_DIFF_INDIRECT]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_DIFFUSE_INDIRECT);
if(out[RRES_OUT_DIFF_COLOR]->hasoutput)
out[RRES_OUT_DIFF_COLOR]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_DIFFUSE_COLOR);
if(out[RRES_OUT_GLOSSY_DIRECT]->hasoutput)
out[RRES_OUT_GLOSSY_DIRECT]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_GLOSSY_DIRECT);
if(out[RRES_OUT_GLOSSY_INDIRECT]->hasoutput)
out[RRES_OUT_GLOSSY_INDIRECT]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_GLOSSY_INDIRECT);
if(out[RRES_OUT_GLOSSY_COLOR]->hasoutput)
out[RRES_OUT_GLOSSY_COLOR]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_GLOSSY_COLOR);
if(out[RRES_OUT_TRANSM_DIRECT]->hasoutput)
out[RRES_OUT_TRANSM_DIRECT]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_TRANSM_DIRECT);
if(out[RRES_OUT_TRANSM_INDIRECT]->hasoutput)
out[RRES_OUT_TRANSM_INDIRECT]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_TRANSM_INDIRECT);
if(out[RRES_OUT_TRANSM_COLOR]->hasoutput)
out[RRES_OUT_TRANSM_COLOR]->data= compbuf_multilayer_get(rd, rl, ima, iuser, SCE_PASS_TRANSM_COLOR);
} }
@@ -402,6 +429,24 @@ static void node_composit_rlayers_out(RenderData *rd, RenderLayer *rl, bNodeStac
out[RRES_OUT_EMIT]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_EMIT); out[RRES_OUT_EMIT]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_EMIT);
if(out[RRES_OUT_ENV]->hasoutput) if(out[RRES_OUT_ENV]->hasoutput)
out[RRES_OUT_ENV]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_ENVIRONMENT); out[RRES_OUT_ENV]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_ENVIRONMENT);
if(out[RRES_OUT_DIFF_DIRECT]->hasoutput)
out[RRES_OUT_DIFF_DIRECT]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_DIFFUSE_DIRECT);
if(out[RRES_OUT_DIFF_INDIRECT]->hasoutput)
out[RRES_OUT_DIFF_INDIRECT]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_DIFFUSE_INDIRECT);
if(out[RRES_OUT_DIFF_COLOR]->hasoutput)
out[RRES_OUT_DIFF_COLOR]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_DIFFUSE_COLOR);
if(out[RRES_OUT_GLOSSY_DIRECT]->hasoutput)
out[RRES_OUT_GLOSSY_DIRECT]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_GLOSSY_DIRECT);
if(out[RRES_OUT_GLOSSY_INDIRECT]->hasoutput)
out[RRES_OUT_GLOSSY_INDIRECT]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_GLOSSY_INDIRECT);
if(out[RRES_OUT_GLOSSY_COLOR]->hasoutput)
out[RRES_OUT_GLOSSY_COLOR]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_GLOSSY_COLOR);
if(out[RRES_OUT_TRANSM_DIRECT]->hasoutput)
out[RRES_OUT_TRANSM_DIRECT]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_TRANSM_DIRECT);
if(out[RRES_OUT_TRANSM_INDIRECT]->hasoutput)
out[RRES_OUT_TRANSM_INDIRECT]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_TRANSM_INDIRECT);
if(out[RRES_OUT_TRANSM_COLOR]->hasoutput)
out[RRES_OUT_TRANSM_COLOR]->data= compbuf_from_pass(rd, rl, rectx, recty, SCE_PASS_TRANSM_COLOR);
} }
static void node_composit_exec_rlayers(void *data, bNode *node, bNodeStack **UNUSED(in), bNodeStack **out) static void node_composit_exec_rlayers(void *data, bNode *node, bNodeStack **UNUSED(in), bNodeStack **out)

100
source/blender/render/intern/source/render_result.c
View File

@@ -224,6 +224,60 @@ static const char *get_pass_name(int passtype, int channel)
if(channel==1) return "Rayhits.G"; if(channel==1) return "Rayhits.G";
return "Rayhits.B"; return "Rayhits.B";
} }
if(passtype == SCE_PASS_DIFFUSE_DIRECT) {
if(channel==-1) return "DiffDir";
if(channel==0) return "DiffDir.R";
if(channel==1) return "DiffDir.G";
return "DiffDir.B";
}
if(passtype == SCE_PASS_DIFFUSE_INDIRECT) {
if(channel==-1) return "DiffInd";
if(channel==0) return "DiffInd.R";
if(channel==1) return "DiffInd.G";
return "DiffInd.B";
}
if(passtype == SCE_PASS_DIFFUSE_COLOR) {
if(channel==-1) return "DiffCol";
if(channel==0) return "DiffCol.R";
if(channel==1) return "DiffCol.G";
return "DiffCol.B";
}
if(passtype == SCE_PASS_GLOSSY_DIRECT) {
if(channel==-1) return "GlossDir";
if(channel==0) return "GlossDir.R";
if(channel==1) return "GlossDir.G";
return "GlossDir.B";
}
if(passtype == SCE_PASS_GLOSSY_INDIRECT) {
if(channel==-1) return "GlossInd";
if(channel==0) return "GlossInd.R";
if(channel==1) return "GlossInd.G";
return "GlossInd.B";
}
if(passtype == SCE_PASS_GLOSSY_COLOR) {
if(channel==-1) return "GlossCol";
if(channel==0) return "GlossCol.R";
if(channel==1) return "GlossCol.G";
return "GlossCol.B";
}
if(passtype == SCE_PASS_TRANSM_DIRECT) {
if(channel==-1) return "TransDir";
if(channel==0) return "TransDir.R";
if(channel==1) return "TransDir.G";
return "TransDir.B";
}
if(passtype == SCE_PASS_TRANSM_INDIRECT) {
if(channel==-1) return "TransInd";
if(channel==0) return "TransInd.R";
if(channel==1) return "TransInd.G";
return "TransInd.B";
}
if(passtype == SCE_PASS_TRANSM_COLOR) {
if(channel==-1) return "TransCol";
if(channel==0) return "TransCol.R";
if(channel==1) return "TransCol.G";
return "TransCol.B";
}
return "Unknown"; return "Unknown";
} }
@@ -286,6 +340,34 @@ static int passtype_from_name(const char *str)
if(strcmp(str, "RayHits")==0) if(strcmp(str, "RayHits")==0)
return SCE_PASS_RAYHITS; return SCE_PASS_RAYHITS;
if(strcmp(str, "DiffDir")==0)
return SCE_PASS_DIFFUSE_DIRECT;
if(strcmp(str, "DiffInd")==0)
return SCE_PASS_DIFFUSE_INDIRECT;
if(strcmp(str, "DiffCol")==0)
return SCE_PASS_DIFFUSE_COLOR;
if(strcmp(str, "GlossDir")==0)
return SCE_PASS_GLOSSY_DIRECT;
if(strcmp(str, "GlossInd")==0)
return SCE_PASS_GLOSSY_INDIRECT;
if(strcmp(str, "GlossCol")==0)
return SCE_PASS_GLOSSY_COLOR;
if(strcmp(str, "TransDir")==0)
return SCE_PASS_TRANSM_DIRECT;
if(strcmp(str, "TransInd")==0)
return SCE_PASS_TRANSM_INDIRECT;
if(strcmp(str, "TransCol")==0)
return SCE_PASS_TRANSM_COLOR;
return 0; return 0;
} }
@@ -430,6 +512,24 @@ RenderResult *render_result_new(Render *re, rcti *partrct, int crop, int savebuf
render_layer_add_pass(rr, rl, 1, SCE_PASS_MIST); render_layer_add_pass(rr, rl, 1, SCE_PASS_MIST);
if(rl->passflag & SCE_PASS_RAYHITS) if(rl->passflag & SCE_PASS_RAYHITS)
render_layer_add_pass(rr, rl, 4, SCE_PASS_RAYHITS); render_layer_add_pass(rr, rl, 4, SCE_PASS_RAYHITS);
if(srl->passflag & SCE_PASS_DIFFUSE_DIRECT)
render_layer_add_pass(rr, rl, 3, SCE_PASS_DIFFUSE_DIRECT);
if(srl->passflag & SCE_PASS_DIFFUSE_INDIRECT)
render_layer_add_pass(rr, rl, 3, SCE_PASS_DIFFUSE_INDIRECT);
if(srl->passflag & SCE_PASS_DIFFUSE_COLOR)
render_layer_add_pass(rr, rl, 3, SCE_PASS_DIFFUSE_COLOR);
if(srl->passflag & SCE_PASS_GLOSSY_DIRECT)
render_layer_add_pass(rr, rl, 3, SCE_PASS_GLOSSY_DIRECT);
if(srl->passflag & SCE_PASS_GLOSSY_INDIRECT)
render_layer_add_pass(rr, rl, 3, SCE_PASS_GLOSSY_INDIRECT);
if(srl->passflag & SCE_PASS_GLOSSY_COLOR)
render_layer_add_pass(rr, rl, 3, SCE_PASS_GLOSSY_COLOR);
if(srl->passflag & SCE_PASS_TRANSM_DIRECT)
render_layer_add_pass(rr, rl, 3, SCE_PASS_TRANSM_DIRECT);
if(srl->passflag & SCE_PASS_TRANSM_INDIRECT)
render_layer_add_pass(rr, rl, 3, SCE_PASS_TRANSM_INDIRECT);
if(srl->passflag & SCE_PASS_TRANSM_COLOR)
render_layer_add_pass(rr, rl, 3, SCE_PASS_TRANSM_COLOR);
} }
/* sss, previewrender and envmap don't do layers, so we make a default one */ /* sss, previewrender and envmap don't do layers, so we make a default one */