Some fixes for CUDA runtime compile:
* When Baking wasn't used we got an error. * On top of Volume Nodes (NODES_FEATURE_VOLUME), we now also check if we need volume sampling code, so we can disable that as well and save some further compilation time.
This commit is contained in:
parent
734d1aec3f
commit
4422b3f919
|
@ -103,6 +103,9 @@ public:
|
|||
/* Use subsurface scattering materials. */
|
||||
bool use_subsurface;
|
||||
|
||||
/* Use volume materials. */
|
||||
bool use_volume;
|
||||
|
||||
/* Use branched integrator. */
|
||||
bool use_integrator_branched;
|
||||
|
||||
|
@ -118,6 +121,7 @@ public:
|
|||
use_camera_motion = false;
|
||||
use_baking = false;
|
||||
use_subsurface = false;
|
||||
use_volume = false;
|
||||
use_integrator_branched = false;
|
||||
}
|
||||
|
||||
|
@ -132,6 +136,7 @@ public:
|
|||
use_camera_motion == requested_features.use_camera_motion &&
|
||||
use_baking == requested_features.use_baking &&
|
||||
use_subsurface == requested_features.use_subsurface &&
|
||||
use_volume == requested_features.use_volume &&
|
||||
use_integrator_branched == requested_features.use_integrator_branched);
|
||||
}
|
||||
|
||||
|
@ -161,6 +166,9 @@ public:
|
|||
if(!use_baking) {
|
||||
build_options += " -D__NO_BAKING__";
|
||||
}
|
||||
if(!use_volume) {
|
||||
build_options += " -D__NO_VOLUME__";
|
||||
}
|
||||
if(!use_subsurface) {
|
||||
build_options += " -D__NO_SUBSURFACE__";
|
||||
}
|
||||
|
|
|
@ -16,7 +16,7 @@
|
|||
|
||||
CCL_NAMESPACE_BEGIN
|
||||
|
||||
#ifndef __NO_BAKING__
|
||||
#ifdef __BAKING__
|
||||
|
||||
ccl_device void compute_light_pass(KernelGlobals *kg, ShaderData *sd, PathRadiance *L, RNG rng,
|
||||
int pass_filter, int sample)
|
||||
|
@ -483,7 +483,7 @@ ccl_device void kernel_bake_evaluate(KernelGlobals *kg, ccl_global uint4 *input,
|
|||
output[i] += make_float4(out.x, out.y, out.z, 1.0f) * output_fac;
|
||||
}
|
||||
|
||||
#endif /* __NO_BAKING__ */
|
||||
#endif /* __BAKING__ */
|
||||
|
||||
ccl_device void kernel_shader_evaluate(KernelGlobals *kg,
|
||||
ccl_global uint4 *input,
|
||||
|
|
|
@ -120,6 +120,7 @@ CCL_NAMESPACE_BEGIN
|
|||
# define __CAMERA_MOTION__
|
||||
# define __OBJECT_MOTION__
|
||||
# define __HAIR__
|
||||
# define __BAKING__
|
||||
# ifdef __KERNEL_EXPERIMENTAL__
|
||||
# define __TRANSPARENT_SHADOWS__
|
||||
# endif
|
||||
|
@ -167,13 +168,14 @@ CCL_NAMESPACE_BEGIN
|
|||
# define __CAMERA_MOTION__
|
||||
# define __OBJECT_MOTION__
|
||||
# define __HAIR__
|
||||
# define __BAKING__
|
||||
#endif
|
||||
|
||||
#ifdef WITH_CYCLES_DEBUG
|
||||
# define __KERNEL_DEBUG__
|
||||
#endif
|
||||
|
||||
/* Scene-based selective featrues compilation. */
|
||||
/* Scene-based selective features compilation. */
|
||||
#ifdef __NO_CAMERA_MOTION__
|
||||
# undef __CAMERA_MOTION__
|
||||
#endif
|
||||
|
@ -183,9 +185,16 @@ CCL_NAMESPACE_BEGIN
|
|||
#ifdef __NO_HAIR__
|
||||
# undef __HAIR__
|
||||
#endif
|
||||
#ifdef __NO_VOLUME__
|
||||
# undef __VOLUME__
|
||||
# undef __VOLUME_SCATTER__
|
||||
#endif
|
||||
#ifdef __NO_SUBSURFACE__
|
||||
# undef __SUBSURFACE__
|
||||
#endif
|
||||
#ifdef __NO_BAKING__
|
||||
# undef __BAKING__
|
||||
#endif
|
||||
#ifdef __NO_BRANCHED_PATH__
|
||||
# undef __BRANCHED_PATH__
|
||||
#endif
|
||||
|
|
|
@ -193,6 +193,7 @@ kernel_cuda_shader(uint4 *input,
|
|||
}
|
||||
}
|
||||
|
||||
#ifdef __BAKING__
|
||||
extern "C" __global__ void
|
||||
CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS)
|
||||
kernel_cuda_bake(uint4 *input, float4 *output, int type, int filter, int sx, int sw, int offset, int sample)
|
||||
|
@ -202,6 +203,7 @@ kernel_cuda_bake(uint4 *input, float4 *output, int type, int filter, int sx, int
|
|||
if(x < sx + sw)
|
||||
kernel_bake_evaluate(NULL, input, output, (ShaderEvalType)type, filter, x, offset, sample);
|
||||
}
|
||||
#endif
|
||||
|
||||
#endif
|
||||
|
||||
|
|
|
@ -528,6 +528,10 @@ void ShaderManager::get_requested_features(Scene *scene,
|
|||
if(output_node->input("Displacement")->link != NULL) {
|
||||
requested_features->nodes_features |= NODE_FEATURE_BUMP;
|
||||
}
|
||||
/* On top of volume nodes, also check if we need volume sampling because
|
||||
* e.g. an Emission node would slip through the NODE_FEATURE_VOLUME check */
|
||||
if(shader->has_volume)
|
||||
requested_features->use_volume |= true;
|
||||
}
|
||||
}
|
||||
|
||||
|
|
Loading…
Reference in New Issue