17 #ifndef __KERNEL_WORK_STEALING_H__
18 #define __KERNEL_WORK_STEALING_H__
28 uint global_work_index,
33 #ifdef __KERNEL_CUDA__
35 uint sample_offset = global_work_index % tile->num_samples;
36 uint pixel_offset = global_work_index / tile->num_samples;
38 uint tile_pixels = tile->w * tile->h;
39 uint sample_offset = global_work_index / tile_pixels;
40 uint pixel_offset = global_work_index - sample_offset * tile_pixels;
42 uint y_offset = pixel_offset / tile->w;
43 uint x_offset = pixel_offset - y_offset * tile->w;
45 *
x = tile->x + x_offset;
46 *
y = tile->y + y_offset;
47 *
sample = tile->start_sample + sample_offset;
50 #ifdef __KERNEL_OPENCL__
51 # pragma OPENCL EXTENSION cl_khr_global_int32_base_atomics : enable
54 #ifdef __SPLIT_KERNEL__
64 if (ray_index >= total_work_size) {
81 return (*global_work_index < total_work_size);
90 bool got_work =
false;
93 got_work = get_next_work_item(
kg,
work_pools, total_work_size, ray_index, global_work_index);
98 uint buffer_offset = (tile->offset +
x +
y * tile->stride) *
kernel_data.film.pass_stride;
102 if ((*aux).w == 0.0f) {
109 got_work = get_next_work_item(
kg,
work_pools, total_work_size, ray_index, global_work_index);
_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 y
#define kernel_assert(cond)
#define ccl_global_size(d)
#define ccl_device_inline
#define CCL_NAMESPACE_END
#define kernel_split_params
__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 ccl_global unsigned int * work_pools
CCL_NAMESPACE_BEGIN ccl_device_inline void get_work_pixel(ccl_global const WorkTile *tile, uint global_work_index, ccl_private uint *x, ccl_private uint *y, ccl_private uint *sample)
static void sample(SocketReader *reader, int x, int y, float color[4])
#define atomic_fetch_and_inc_uint32(p)