Blender  V2.93
kernel_data_init.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 
18 
19 /* This kernel Initializes structures needed in path-iteration kernels.
20  *
21  * Note on Queues:
22  * All slots in queues are initialized to queue empty slot;
23  * The number of elements in the queues is initialized to 0;
24  */
25 
26 #ifndef __KERNEL_CPU__
28 #else
30 #endif
31  KernelGlobals *kg,
34  int num_elements,
35  ccl_global char *ray_state,
36 
37 #ifdef __KERNEL_OPENCL__
39 #endif
40 
41  int start_sample,
42  int end_sample,
43  int sx,
44  int sy,
45  int sw,
46  int sh,
47  int offset,
48  int stride,
49  ccl_global int *Queue_index, /* Tracks the number of elements in queues */
50  int queuesize, /* size (capacity) of the queue */
51  ccl_global char *use_queues_flag, /* flag to decide if scene-intersect kernel should use queues
52  to fetch ray index */
53  ccl_global unsigned int *work_pools, /* Work pool for each work group */
54  unsigned int num_samples,
55  ccl_global float *buffer)
56 {
57 #ifdef KERNEL_STUB
58  STUB_ASSERT(KERNEL_ARCH, data_init);
59 #else
60 
61 # ifdef __KERNEL_OPENCL__
62  kg->data = data;
63 # endif
64 
65  kernel_split_params.tile.x = sx;
66  kernel_split_params.tile.y = sy;
67  kernel_split_params.tile.w = sw;
68  kernel_split_params.tile.h = sh;
69 
70  kernel_split_params.tile.start_sample = start_sample;
71  kernel_split_params.tile.num_samples = num_samples;
72 
73  kernel_split_params.tile.offset = offset;
74  kernel_split_params.tile.stride = stride;
75 
76  kernel_split_params.tile.buffer = buffer;
77 
78  kernel_split_params.total_work_size = sw * sh * num_samples;
79 
80  kernel_split_params.work_pools = work_pools;
81 
82  kernel_split_params.queue_index = Queue_index;
83  kernel_split_params.queue_size = queuesize;
84  kernel_split_params.use_queues_flag = use_queues_flag;
85 
87 
88 # ifdef __KERNEL_OPENCL__
89  kernel_set_buffer_pointers(kg, KERNEL_BUFFER_ARGS);
90  kernel_set_buffer_info(kg);
91 # endif
92 
93  int thread_index = ccl_global_id(1) * ccl_global_size(0) + ccl_global_id(0);
94 
95  /* Initialize queue data and queue index. */
96  if (thread_index < queuesize) {
97  for (int i = 0; i < NUM_QUEUES; i++) {
98  kernel_split_state.queue_data[i * queuesize + thread_index] = QUEUE_EMPTY_SLOT;
99  }
100  }
101 
102  if (thread_index == 0) {
103  for (int i = 0; i < NUM_QUEUES; i++) {
104  Queue_index[i] = 0;
105  }
106 
107  /* The scene-intersect kernel should not use the queues very first time.
108  * since the queue would be empty.
109  */
110  *use_queues_flag = 0;
111  }
112 #endif /* KERENL_STUB */
113 }
114 
_GL_VOID GLfloat value _GL_VOID_RET _GL_VOID const GLuint GLboolean *residences _GL_BOOL_RET _GL_VOID GLsizei GLfloat GLfloat GLfloat GLfloat const GLubyte *bitmap _GL_VOID_RET _GL_VOID GLenum const void *lists _GL_VOID_RET _GL_VOID const GLdouble *equation _GL_VOID_RET _GL_VOID GLdouble GLdouble blue _GL_VOID_RET _GL_VOID GLfloat GLfloat blue _GL_VOID_RET _GL_VOID GLint GLint blue _GL_VOID_RET _GL_VOID GLshort GLshort blue _GL_VOID_RET _GL_VOID GLubyte GLubyte blue _GL_VOID_RET _GL_VOID GLuint GLuint blue _GL_VOID_RET _GL_VOID GLushort GLushort blue _GL_VOID_RET _GL_VOID GLbyte GLbyte GLbyte alpha _GL_VOID_RET _GL_VOID GLdouble GLdouble GLdouble alpha _GL_VOID_RET _GL_VOID GLfloat GLfloat GLfloat alpha _GL_VOID_RET _GL_VOID GLint GLint GLint alpha _GL_VOID_RET _GL_VOID GLshort GLshort GLshort alpha _GL_VOID_RET _GL_VOID GLubyte GLubyte GLubyte alpha _GL_VOID_RET _GL_VOID GLuint GLuint GLuint alpha _GL_VOID_RET _GL_VOID GLushort GLushort GLushort alpha _GL_VOID_RET _GL_VOID GLenum mode _GL_VOID_RET _GL_VOID GLint GLsizei GLsizei GLenum type _GL_VOID_RET _GL_VOID GLsizei GLenum GLenum const void *pixels _GL_VOID_RET _GL_VOID const void *pointer _GL_VOID_RET _GL_VOID GLdouble v _GL_VOID_RET _GL_VOID GLfloat v _GL_VOID_RET _GL_VOID GLint GLint i2 _GL_VOID_RET _GL_VOID GLint j _GL_VOID_RET _GL_VOID GLfloat param _GL_VOID_RET _GL_VOID GLint param _GL_VOID_RET _GL_VOID GLdouble GLdouble GLdouble GLdouble GLdouble zFar _GL_VOID_RET _GL_UINT GLdouble *equation _GL_VOID_RET _GL_VOID GLenum GLint *params _GL_VOID_RET _GL_VOID GLenum GLfloat *v _GL_VOID_RET _GL_VOID GLenum GLfloat *params _GL_VOID_RET _GL_VOID GLfloat *values _GL_VOID_RET _GL_VOID GLushort *values _GL_VOID_RET _GL_VOID GLenum GLfloat *params _GL_VOID_RET _GL_VOID GLenum GLdouble *params _GL_VOID_RET _GL_VOID GLenum GLint *params _GL_VOID_RET _GL_VOID GLsizei stride
#define KERNEL_FUNCTION_FULL_NAME(name)
Definition: filter.h:30
#define KERNEL_ARCH
Definition: filter.h:47
#define ccl_global_id(d)
#define ccl_global_size(d)
#define ccl_device
#define ccl_constant
#define ccl_global
#define CCL_NAMESPACE_END
#define __KERNEL_OPENCL__
void KERNEL_FUNCTION_FULL_NAME() data_init(KernelGlobals *kg, ccl_constant KernelData *data, ccl_global void *split_data_buffer, int num_elements, ccl_global char *ray_state, int start_sample, int end_sample, int sx, int sy, int sw, int sh, int offset, int stride, ccl_global int *Queue_index, int queuesize, ccl_global char *use_queues_flag, ccl_global unsigned int *work_pool_wgs, unsigned int num_samples, ccl_global float *buffer)
CCL_NAMESPACE_BEGIN ccl_device void kernel_data_init(KernelGlobals *kg, ccl_constant KernelData *data, ccl_global void *split_data_buffer, int num_elements, ccl_global char *ray_state, int start_sample, int end_sample, int sx, int sy, int sw, int sh, int offset, int stride, ccl_global int *Queue_index, int queuesize, ccl_global char *use_queues_flag, ccl_global unsigned int *work_pools, unsigned int num_samples, ccl_global float *buffer)
ccl_device_inline void split_data_init(KernelGlobals *kg, ccl_global SplitData *split_data, size_t num_elements, ccl_global void *data, ccl_global char *ray_state)
#define kernel_split_params
#define kernel_split_state
__kernel void ccl_constant KernelData ccl_global void ccl_global char ccl_global int ccl_global char ccl_global unsigned int ccl_global float * buffer
__kernel void ccl_constant KernelData ccl_global void ccl_global char ccl_global int ccl_global char * use_queues_flag
kernel_set_buffer_pointers(kg, KERNEL_BUFFER_ARGS)
__kernel void ccl_constant KernelData ccl_global void * split_data_buffer
__kernel void ccl_constant KernelData ccl_global void ccl_global char KERNEL_BUFFER_PARAMS
__kernel void ccl_constant KernelData ccl_global void ccl_global char * ray_state
__kernel void ccl_constant KernelData ccl_global void ccl_global char ccl_global int ccl_global char ccl_global unsigned int * work_pools
#define QUEUE_EMPTY_SLOT
@ NUM_QUEUES