Blender  V2.93
kernel_split_common.h
Go to the documentation of this file.
1 /*
2  * Copyright 2011-2015 Blender Foundation
3  *
4  * Licensed under the Apache License, Version 2.0 (the "License");
5  * you may not use this file except in compliance with the License.
6  * You may obtain a copy of the License at
7  *
8  * http://www.apache.org/licenses/LICENSE-2.0
9  *
10  * Unless required by applicable law or agreed to in writing, software
11  * distributed under the License is distributed on an "AS IS" BASIS,
12  * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
13  * See the License for the specific language governing permissions and
14  * limitations under the License.
15  */
16 
17 #ifndef __KERNEL_SPLIT_H__
18 #define __KERNEL_SPLIT_H__
19 
20 // clang-format off
21 #include "kernel/kernel_math.h"
22 #include "kernel/kernel_types.h"
23 
25 
26 #include "kernel/kernel_globals.h"
27 #include "kernel/kernel_color.h"
28 
29 #ifdef __OSL__
30 # include "kernel/osl/osl_shader.h"
31 #endif
32 
33 #ifdef __KERNEL_OPENCL__
35 #endif
36 #ifdef __KERNEL_CUDA__
38 #endif
39 #ifdef __KERNEL_CPU__
41 #endif
42 
43 #include "util/util_atomic.h"
44 
45 #include "kernel/kernel_path.h"
46 #ifdef __BRANCHED_PATH__
48 #endif
49 
50 #include "kernel/kernel_queues.h"
52 
53 #ifdef __BRANCHED_PATH__
55 #endif
56 // clang-format on
57 
59 
60 ccl_device_inline void kernel_split_path_end(KernelGlobals *kg, int ray_index)
61 {
62  ccl_global char *ray_state = kernel_split_state.ray_state;
63 
64 #ifdef __BRANCHED_PATH__
65 # ifdef __SUBSURFACE__
66  ccl_addr_space SubsurfaceIndirectRays *ss_indirect = &kernel_split_state.ss_rays[ray_index];
67 
68  if (ss_indirect->num_rays) {
70  }
71  else
72 # endif /* __SUBSURFACE__ */
74  int orig_ray = kernel_split_state.branched_state[ray_index].original_ray;
75 
76  PathRadiance *L = &kernel_split_state.path_radiance[ray_index];
77  PathRadiance *orig_ray_L = &kernel_split_state.path_radiance[orig_ray];
78 
80  path_radiance_accum_sample(orig_ray_L, L);
81 
83  (ccl_global uint *)&kernel_split_state.branched_state[orig_ray].shared_sample_count);
84 
86  }
87  else if (IS_FLAG(ray_state, ray_index, RAY_BRANCHED_LIGHT_INDIRECT)) {
89  }
90  else if (IS_FLAG(ray_state, ray_index, RAY_BRANCHED_VOLUME_INDIRECT)) {
92  }
93  else if (IS_FLAG(ray_state, ray_index, RAY_BRANCHED_SUBSURFACE_INDIRECT)) {
95  }
96  else {
98  }
99 #else
101 #endif
102 }
103 
105 
106 #endif /* __KERNEL_SPLIT_H__ */
unsigned int uint
Definition: BLI_sys_types.h:83
ccl_device_inline void path_radiance_sum_indirect(PathRadiance *L)
ccl_device_inline void path_radiance_accum_sample(PathRadiance *L, PathRadiance *L_sample)
#define ccl_addr_space
#define ccl_device_inline
#define ccl_global
#define CCL_NAMESPACE_END
CCL_NAMESPACE_BEGIN ccl_device_inline void kernel_split_path_end(KernelGlobals *kg, int ray_index)
#define kernel_split_state
__kernel void ccl_constant KernelData ccl_global void ccl_global char * ray_state
#define IS_FLAG(ray_state, ray_index, flag)
#define ASSIGN_RAY_STATE(ray_state, ray_index, state)
@ RAY_UPDATE_BUFFER
@ RAY_BRANCHED_VOLUME_INDIRECT
@ RAY_LIGHT_INDIRECT_NEXT_ITER
@ RAY_BRANCHED_SUBSURFACE_INDIRECT
@ RAY_BRANCHED_LIGHT_INDIRECT
@ RAY_BRANCHED_INDIRECT_SHARED
@ RAY_SUBSURFACE_INDIRECT_NEXT_ITER
@ RAY_VOLUME_INDIRECT_NEXT_ITER
@ RAY_INACTIVE
#define L
#define atomic_fetch_and_dec_uint32(p)
Definition: util_atomic.h:30