From 49ab696f914b6ffa205c077a6484882514733237 Mon Sep 17 00:00:00 2001 From: Andreas Kloeckner Date: Mon, 6 Aug 2018 12:11:52 -0500 Subject: [PATCH] Delete cffi bits --- .gitmodules | 6 +- cffi_build.py.in | 86 ---- cl_gl_types.h | 12 - cl_types.h | 128 ----- src/c_wrapper/bitlog.cpp | 59 --- src/c_wrapper/buffer.cpp | 235 --------- src/c_wrapper/buffer.h | 27 - src/c_wrapper/clhelper.h | 254 ---------- src/c_wrapper/clinfo_ext.h | 129 ----- src/c_wrapper/clobj.h | 149 ------ src/c_wrapper/command_queue.cpp | 132 ----- src/c_wrapper/command_queue.h | 64 --- src/c_wrapper/context.cpp | 153 ------ src/c_wrapper/context.h | 34 -- src/c_wrapper/debug.cpp | 84 ---- src/c_wrapper/debug.h | 33 -- src/c_wrapper/device.cpp | 375 -------------- src/c_wrapper/device.h | 61 --- src/c_wrapper/error.h | 337 ------------- src/c_wrapper/event.cpp | 294 ----------- src/c_wrapper/event.h | 87 ---- src/c_wrapper/function.h | 121 ----- src/c_wrapper/gl_obj.cpp | 155 ------ src/c_wrapper/gl_obj.h | 46 -- src/c_wrapper/image.cpp | 237 --------- src/c_wrapper/image.h | 50 -- src/c_wrapper/kernel.cpp | 213 -------- src/c_wrapper/kernel.h | 44 -- src/c_wrapper/memory_map.cpp | 115 ----- src/c_wrapper/memory_map.h | 37 -- src/c_wrapper/memory_object.cpp | 116 ----- src/c_wrapper/memory_object.h | 56 --- src/c_wrapper/mingw-std-threads | 1 - src/c_wrapper/platform.cpp | 109 ---- src/c_wrapper/platform.h | 27 - src/c_wrapper/program.cpp | 269 ---------- src/c_wrapper/program.h | 58 --- src/c_wrapper/pyhelper.cpp | 18 - src/c_wrapper/pyhelper.h | 43 -- src/c_wrapper/pyopencl_ext.h | 58 --- src/c_wrapper/sampler.cpp | 54 -- src/c_wrapper/sampler.h | 33 -- src/c_wrapper/svm.cpp | 173 ------- src/c_wrapper/svm.h | 4 - src/c_wrapper/utils.h | 551 -------------------- src/c_wrapper/wrap_cl.cpp | 123 ----- src/c_wrapper/wrap_cl.h | 171 ------- src/c_wrapper/wrap_cl_core.h | 399 --------------- src/c_wrapper/wrap_cl_gl_core.h | 18 - src/c_wrapper/wrap_constants.cpp | 827 ------------------------------- 50 files changed, 3 insertions(+), 6862 deletions(-) delete mode 100644 cffi_build.py.in delete mode 100644 cl_gl_types.h delete mode 100644 cl_types.h delete mode 100644 src/c_wrapper/bitlog.cpp delete mode 100644 src/c_wrapper/buffer.cpp delete mode 100644 src/c_wrapper/buffer.h delete mode 100644 src/c_wrapper/clhelper.h delete mode 100644 src/c_wrapper/clinfo_ext.h delete mode 100644 src/c_wrapper/clobj.h delete mode 100644 src/c_wrapper/command_queue.cpp delete mode 100644 src/c_wrapper/command_queue.h delete mode 100644 src/c_wrapper/context.cpp delete mode 100644 src/c_wrapper/context.h delete mode 100644 src/c_wrapper/debug.cpp delete mode 100644 src/c_wrapper/debug.h delete mode 100644 src/c_wrapper/device.cpp delete mode 100644 src/c_wrapper/device.h delete mode 100644 src/c_wrapper/error.h delete mode 100644 src/c_wrapper/event.cpp delete mode 100644 src/c_wrapper/event.h delete mode 100644 src/c_wrapper/function.h delete mode 100644 src/c_wrapper/gl_obj.cpp delete mode 100644 src/c_wrapper/gl_obj.h delete mode 100644 src/c_wrapper/image.cpp delete mode 100644 src/c_wrapper/image.h delete mode 100644 src/c_wrapper/kernel.cpp delete mode 100644 src/c_wrapper/kernel.h delete mode 100644 src/c_wrapper/memory_map.cpp delete mode 100644 src/c_wrapper/memory_map.h delete mode 100644 src/c_wrapper/memory_object.cpp delete mode 100644 src/c_wrapper/memory_object.h delete mode 160000 src/c_wrapper/mingw-std-threads delete mode 100644 src/c_wrapper/platform.cpp delete mode 100644 src/c_wrapper/platform.h delete mode 100644 src/c_wrapper/program.cpp delete mode 100644 src/c_wrapper/program.h delete mode 100644 src/c_wrapper/pyhelper.cpp delete mode 100644 src/c_wrapper/pyhelper.h delete mode 100644 src/c_wrapper/pyopencl_ext.h delete mode 100644 src/c_wrapper/sampler.cpp delete mode 100644 src/c_wrapper/sampler.h delete mode 100644 src/c_wrapper/svm.cpp delete mode 100644 src/c_wrapper/svm.h delete mode 100644 src/c_wrapper/utils.h delete mode 100644 src/c_wrapper/wrap_cl.cpp delete mode 100644 src/c_wrapper/wrap_cl.h delete mode 100644 src/c_wrapper/wrap_cl_core.h delete mode 100644 src/c_wrapper/wrap_cl_gl_core.h delete mode 100644 src/c_wrapper/wrap_constants.cpp diff --git a/.gitmodules b/.gitmodules index cb5a4e23..b675a6cc 100644 --- a/.gitmodules +++ b/.gitmodules @@ -1,6 +1,6 @@ [submodule "pyopencl/compyte"] path = pyopencl/compyte url = https://github.com/inducer/compyte -[submodule "src/c_wrapper/mingw-std-threads"] - path = src/c_wrapper/mingw-std-threads - url = https://github.com/meganz/mingw-std-threads.git +[submodule "pybind11"] + path = pybind11 + url = https://github.com/pybind/pybind11.git diff --git a/cffi_build.py.in b/cffi_build.py.in deleted file mode 100644 index f948c824..00000000 --- a/cffi_build.py.in +++ /dev/null @@ -1,86 +0,0 @@ -from __future__ import absolute_import, print_function - -__copyright__ = """ -Copyright (C) 2009-15 Andreas Kloeckner -Copyright (C) 2013 Marko Bencun -Copyright (C) 2014 Yuyi Chao -""" - -__license__ = """ -Permission is hereby granted, free of charge, to any person obtaining a copy -of this software and associated documentation files (the "Software"), to deal -in the Software without restriction, including without limitation the rights -to use, copy, modify, merge, publish, distribute, sublicense, and/or sell -copies of the Software, and to permit persons to whom the Software is -furnished to do so, subject to the following conditions: - -The above copyright notice and this permission notice shall be included in -all copies or substantial portions of the Software. - -THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR -IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, -FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE -AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER -LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, -OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN -THE SOFTWARE. -""" - - -from cffi import FFI - -ffi = FFI() - - -with open("cl_types.h", "rt") as f: - ffi.cdef(f.read()) - -if {CL_ENABLE_GL}: - with open("cl_gl_types.h") as f: - ffi.cdef(f.read()) - -with open("src/c_wrapper/wrap_cl_core.h", "rt") as f: - ffi.cdef(f.read()) - -if {CL_ENABLE_GL}: - with open("src/c_wrapper/wrap_cl_gl_core.h") as f: - ffi.cdef(f.read()) - -ffi.set_source("pyopencl._cffi", - """ - #include "wrap_cl.h" - """, - define_macros=list({EXTRA_DEFINES}.items()), - include_dirs=( - {CL_INC_DIR} + ["src/c_wrapper/"]), - library_dirs={CL_LIB_DIR}, - libraries={CL_LIBNAME}, - extra_compile_args=({CXXFLAGS}), - extra_link_args={LDFLAGS}, - source_extension=".cpp", - sources=[ - "src/c_wrapper/wrap_cl.cpp", - "src/c_wrapper/wrap_constants.cpp", - "src/c_wrapper/bitlog.cpp", - "src/c_wrapper/pyhelper.cpp", - "src/c_wrapper/platform.cpp", - "src/c_wrapper/device.cpp", - "src/c_wrapper/context.cpp", - "src/c_wrapper/command_queue.cpp", - "src/c_wrapper/event.cpp", - "src/c_wrapper/memory_object.cpp", - "src/c_wrapper/svm.cpp", - "src/c_wrapper/image.cpp", - "src/c_wrapper/gl_obj.cpp", - "src/c_wrapper/memory_map.cpp", - "src/c_wrapper/buffer.cpp", - "src/c_wrapper/sampler.cpp", - "src/c_wrapper/program.cpp", - "src/c_wrapper/kernel.cpp", - "src/c_wrapper/debug.cpp", - ] - ) - - -if __name__ == "__main__": - ffi.compile() diff --git a/cl_gl_types.h b/cl_gl_types.h deleted file mode 100644 index ea0e7e4d..00000000 --- a/cl_gl_types.h +++ /dev/null @@ -1,12 +0,0 @@ -/* cl_gl.h */ -typedef cl_uint cl_gl_object_type; -typedef cl_uint cl_gl_texture_info; -typedef cl_uint cl_gl_platform_info; -typedef struct __GLsync *cl_GLsync; -typedef cl_uint cl_gl_context_info; - -/* cl_egl.h */ -typedef void* CLeglImageKHR; -typedef void* CLeglDisplayKHR; -typedef void* CLeglSyncKHR; -typedef intptr_t cl_egl_image_properties_khr; diff --git a/cl_types.h b/cl_types.h deleted file mode 100644 index 5df16013..00000000 --- a/cl_types.h +++ /dev/null @@ -1,128 +0,0 @@ -/* gl.h */ -typedef unsigned int GLenum; -typedef int GLint; /* 4-byte signed */ -typedef unsigned int GLuint; /* 4-byte unsigned */ - - -/* cl.h */ -/* scalar types */ -typedef int8_t cl_char; -typedef uint8_t cl_uchar; -typedef int16_t cl_short; -typedef uint16_t cl_ushort; -typedef int32_t cl_int; -typedef uint32_t cl_uint; -typedef int64_t cl_long; -typedef uint64_t cl_ulong; - -typedef uint16_t cl_half; -typedef float cl_float; -typedef double cl_double; - - -typedef struct _cl_platform_id * cl_platform_id; -typedef struct _cl_device_id * cl_device_id; -typedef struct _cl_context * cl_context; -typedef struct _cl_command_queue * cl_command_queue; -typedef struct _cl_mem * cl_mem; -typedef struct _cl_program * cl_program; -typedef struct _cl_kernel * cl_kernel; -typedef struct _cl_event * cl_event; -typedef struct _cl_sampler * cl_sampler; - -/* WARNING! Unlike cl_ types in cl_platform.h, cl_bool is not guaranteed to be -the same size as the bool in kernels. */ -typedef cl_uint cl_bool; -typedef cl_ulong cl_bitfield; -typedef cl_bitfield cl_device_type; -typedef cl_uint cl_platform_info; -typedef cl_uint cl_device_info; -typedef cl_bitfield cl_device_fp_config; -typedef cl_uint cl_device_mem_cache_type; -typedef cl_uint cl_device_local_mem_type; -typedef cl_bitfield cl_device_exec_capabilities; -typedef cl_bitfield cl_device_svm_capabilities; // 2.0 -typedef cl_bitfield cl_command_queue_properties; -typedef intptr_t cl_device_partition_property; -typedef cl_bitfield cl_device_affinity_domain; - -typedef intptr_t cl_context_properties; -typedef cl_uint cl_context_info; -typedef cl_uint cl_command_queue_info; -typedef cl_uint cl_channel_order; -typedef cl_uint cl_channel_type; -typedef cl_bitfield cl_mem_flags; -typedef cl_bitfield cl_svm_mem_flags; // 2.0 -typedef cl_uint cl_mem_object_type; -typedef cl_uint cl_mem_info; -typedef cl_bitfield cl_mem_migration_flags; -typedef cl_uint cl_image_info; -typedef cl_uint cl_buffer_create_type; -typedef cl_uint cl_addressing_mode; -typedef cl_uint cl_filter_mode; -typedef cl_uint cl_sampler_info; -typedef cl_bitfield cl_map_flags; -typedef intptr_t cl_pipe_properties; // 2.0 -typedef cl_uint cl_pipe_info; // 2.0 -typedef cl_uint cl_program_info; -typedef cl_uint cl_program_build_info; -typedef cl_uint cl_program_binary_type; -typedef cl_int cl_build_status; -typedef cl_uint cl_kernel_info; -typedef cl_uint cl_kernel_arg_info; -typedef cl_uint cl_kernel_arg_address_qualifier; -typedef cl_uint cl_kernel_arg_access_qualifier; -typedef cl_bitfield cl_kernel_arg_type_qualifier; -typedef cl_uint cl_kernel_work_group_info; -typedef cl_uint cl_event_info; -typedef cl_uint cl_command_type; -typedef cl_uint cl_profiling_info; -typedef cl_bitfield cl_sampler_properties; // 2.0 -typedef cl_uint cl_kernel_exec_info; // 2.0 - -typedef struct _cl_image_format { - cl_channel_order image_channel_order; - cl_channel_type image_channel_data_type; -} cl_image_format; - -typedef struct _cl_image_desc { - cl_mem_object_type image_type; - size_t image_width; - size_t image_height; - size_t image_depth; - size_t image_array_size; - size_t image_row_pitch; - size_t image_slice_pitch; - cl_uint num_mip_levels; - cl_uint num_samples; - cl_mem buffer; -} cl_image_desc; - -typedef struct _cl_buffer_region { - size_t origin; - size_t size; -} cl_buffer_region; - -/* cl_ext.h */ - -typedef union -{ - struct { cl_uint type; cl_uint data[5]; } raw; - struct { cl_uint type; cl_char unused[17]; cl_char bus; cl_char device; cl_char function; } pcie; -} cl_device_topology_amd; - -/* -typedef cl_ulong cl_device_partition_property_ext; -typedef cl_uint cl_image_pitch_info_qcom; -typedef struct _cl_mem_ext_host_ptr { - cl_uint allocation_type; - cl_uint host_cache_policy; -} cl_mem_ext_host_ptr; -typedef struct _cl_mem_ion_host_ptr { - cl_mem_ext_host_ptr ext_host_ptr; - int ion_filedesc; - void* ion_hostptr; -} cl_mem_ion_host_ptr; - -typedef cl_bitfield cl_mem_migration_flags_ext; -*/ diff --git a/src/c_wrapper/bitlog.cpp b/src/c_wrapper/bitlog.cpp deleted file mode 100644 index 418eb4d8..00000000 --- a/src/c_wrapper/bitlog.cpp +++ /dev/null @@ -1,59 +0,0 @@ -#include "wrap_cl.h" -#include "function.h" - -#include -#include - -/* from http://graphics.stanford.edu/~seander/bithacks.html */ -static const char log_table_8[] = { - 0, 0, 1, 1, 2, 2, 2, 2, 3, 3, 3, 3, 3, 3, 3, 3, - 4, 4, 4, 4, 4, 4, 4, 4, 4, 4, 4, 4, 4, 4, 4, 4, - 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, - 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, 5, - 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, - 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, - 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, - 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, 6, - 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, - 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, - 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, - 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, - 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, - 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, - 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, - 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7, 7 -}; - -static PYOPENCL_INLINE unsigned -bitlog2_16(uint16_t v) -{ - if (unsigned long t = v >> 8) { - return 8 + log_table_8[t]; - } else { - return log_table_8[v]; - } -} - -static PYOPENCL_INLINE unsigned -bitlog2_32(uint32_t v) -{ - if (uint16_t t = v >> 16) { - return 16 + bitlog2_16(t); - } else { - return bitlog2_16(v); - } -} - -unsigned -bitlog2(unsigned long v) -{ -#if (ULONG_MAX != 4294967295) - if (uint32_t t = v >> 32) { - return 32 + bitlog2_32(t); - } else { -#endif - return bitlog2_32(v); -#if (ULONG_MAX != 4294967295) - } -#endif -} diff --git a/src/c_wrapper/buffer.cpp b/src/c_wrapper/buffer.cpp deleted file mode 100644 index 70e1ff3e..00000000 --- a/src/c_wrapper/buffer.cpp +++ /dev/null @@ -1,235 +0,0 @@ -#include -#include "buffer.h" -#include "context.h" -#include "command_queue.h" -#include "event.h" - -template void print_clobj(std::ostream&, const buffer*); - -PYOPENCL_USE_RESULT static PYOPENCL_INLINE buffer* -new_buffer(cl_mem mem) -{ - return pyopencl_convert_obj(buffer, clReleaseMemObject, mem); -} - -#if PYOPENCL_CL_VERSION >= 0x1010 -PYOPENCL_USE_RESULT buffer* -buffer::get_sub_region(size_t orig, size_t size, cl_mem_flags flags) const -{ - cl_buffer_region reg = {orig, size}; - - auto mem = retry_mem_error([&] { - return pyopencl_call_guarded(clCreateSubBuffer, PYOPENCL_CL_CASTABLE_THIS, flags, - CL_BUFFER_CREATE_TYPE_REGION, ®); - }); - return new_buffer(mem); -} - -#endif - -// c wrapper - -// Buffer -error* -create_buffer(clobj_t *buffer, clobj_t _ctx, cl_mem_flags flags, - size_t size, void *hostbuf) -{ - auto ctx = static_cast(_ctx); - return c_handle_retry_mem_error([&] { - auto mem = pyopencl_call_guarded(clCreateBuffer, ctx, - flags, size, hostbuf); - *buffer = new_buffer(mem); - }); -} - -error* -enqueue_read_buffer(clobj_t *evt, clobj_t _queue, clobj_t _mem, - void *buffer, size_t size, size_t device_offset, - const clobj_t *_wait_for, uint32_t num_wait_for, - int block, void *pyobj) -{ - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - auto mem = static_cast(_mem); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueReadBuffer, queue, mem, bool(block), device_offset, - size, buffer, wait_for, nanny_event_out(evt, pyobj)); - }); -} - -error* -enqueue_write_buffer(clobj_t *evt, clobj_t _queue, clobj_t _mem, - const void *buffer, size_t size, size_t device_offset, - const clobj_t *_wait_for, uint32_t num_wait_for, - int block, void *pyobj) -{ - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - auto mem = static_cast(_mem); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueWriteBuffer, queue, mem, bool(block), device_offset, - size, buffer, wait_for, nanny_event_out(evt, pyobj)); - }); -} - -error* -enqueue_copy_buffer(clobj_t *evt, clobj_t _queue, clobj_t _src, clobj_t _dst, - ptrdiff_t byte_count, size_t src_offset, size_t dst_offset, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ - auto queue = static_cast(_queue); - auto src = static_cast(_src); - auto dst = static_cast(_dst); - return c_handle_error([&] { - if (byte_count < 0) { - size_t byte_count_src = 0; - size_t byte_count_dst = 0; - pyopencl_call_guarded( - clGetMemObjectInfo, src, CL_MEM_SIZE, - sizeof(byte_count), &byte_count_src, nullptr); - pyopencl_call_guarded( - clGetMemObjectInfo, src, CL_MEM_SIZE, - sizeof(byte_count), &byte_count_dst, nullptr); - byte_count = std::min(byte_count_src, byte_count_dst); - } - const auto wait_for = buf_from_class(_wait_for, - num_wait_for); - retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueCopyBuffer, queue, src, dst, src_offset, - dst_offset, byte_count, wait_for, event_out(evt)); - }); - }); -} - - -error* -enqueue_fill_buffer(clobj_t *evt, clobj_t _queue, clobj_t _mem, void *pattern, - size_t psize, size_t offset, size_t size, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - auto mem = static_cast(_mem); - // TODO debug print pattern - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueFillBuffer, queue, mem, pattern, - psize, offset, size, wait_for, - event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED(clEnqueueFillBuffer, "CL 1.1 and below") -#endif -} - - -// {{{ rectangular transfers - -error* -enqueue_read_buffer_rect(clobj_t *evt, clobj_t _queue, clobj_t _mem, void *buf, - const size_t *_buf_orig, size_t buf_orig_l, - const size_t *_host_orig, size_t host_orig_l, - const size_t *_reg, size_t reg_l, - const size_t *_buf_pitches, size_t buf_pitches_l, - const size_t *_host_pitches, size_t host_pitches_l, - const clobj_t *_wait_for, uint32_t num_wait_for, - int block, void *pyobj) -{ -#if PYOPENCL_CL_VERSION >= 0x1010 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - auto mem = static_cast(_mem); - ConstBuffer buf_orig(_buf_orig, buf_orig_l); - ConstBuffer host_orig(_host_orig, host_orig_l); - ConstBuffer reg(_reg, reg_l, 1); - ConstBuffer buf_pitches(_buf_pitches, buf_pitches_l); - ConstBuffer host_pitches(_host_pitches, host_pitches_l); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueReadBufferRect, queue, mem, bool(block), buf_orig, - host_orig, reg, buf_pitches[0], buf_pitches[1], host_pitches[0], - host_pitches[1], buf, wait_for, nanny_event_out(evt, pyobj)); - }); -#else - PYOPENCL_UNSUPPORTED(clEnqueueReadBufferRect, "CL 1.0") -#endif -} - -error* -enqueue_write_buffer_rect(clobj_t *evt, clobj_t _queue, clobj_t _mem, void *buf, - const size_t *_buf_orig, size_t buf_orig_l, - const size_t *_host_orig, size_t host_orig_l, - const size_t *_reg, size_t reg_l, - const size_t *_buf_pitches, size_t buf_pitches_l, - const size_t *_host_pitches, size_t host_pitches_l, - const clobj_t *_wait_for, uint32_t num_wait_for, - int block, void *pyobj) -{ -#if PYOPENCL_CL_VERSION >= 0x1010 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - auto mem = static_cast(_mem); - ConstBuffer buf_orig(_buf_orig, buf_orig_l); - ConstBuffer host_orig(_host_orig, host_orig_l); - ConstBuffer reg(_reg, reg_l, 1); - ConstBuffer buf_pitches(_buf_pitches, buf_pitches_l); - ConstBuffer host_pitches(_host_pitches, host_pitches_l); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueWriteBufferRect, queue, mem, bool(block), buf_orig, - host_orig, reg, buf_pitches[0], buf_pitches[1], host_pitches[0], - host_pitches[1], buf, wait_for, nanny_event_out(evt, pyobj)); - }); -#else - PYOPENCL_UNSUPPORTED(clEnqueueWriteBufferRect, "CL 1.0") -#endif -} - -error* -enqueue_copy_buffer_rect(clobj_t *evt, clobj_t _queue, clobj_t _src, - clobj_t _dst, const size_t *_src_orig, - size_t src_orig_l, const size_t *_dst_orig, - size_t dst_orig_l, const size_t *_reg, size_t reg_l, - const size_t *_src_pitches, size_t src_pitches_l, - const size_t *_dst_pitches, size_t dst_pitches_l, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x1010 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - auto src = static_cast(_src); - auto dst = static_cast(_dst); - ConstBuffer src_orig(_src_orig, src_orig_l); - ConstBuffer dst_orig(_dst_orig, dst_orig_l); - ConstBuffer reg(_reg, reg_l, 1); - ConstBuffer src_pitches(_src_pitches, src_pitches_l); - ConstBuffer dst_pitches(_dst_pitches, dst_pitches_l); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueCopyBufferRect, queue, src, dst, src_orig, dst_orig, - reg, src_pitches[0], src_pitches[1], dst_pitches[0], - dst_pitches[1], wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED(clEnqueueCopyBufferRect, "CL 1.0") -#endif -} - -// }}} - -error* -buffer__get_sub_region(clobj_t *_sub_buf, clobj_t _buf, size_t orig, - size_t size, cl_mem_flags flags) -{ -#if PYOPENCL_CL_VERSION >= 0x1010 - auto buf = static_cast(_buf); - return c_handle_error([&] { - *_sub_buf = buf->get_sub_region(orig, size, flags); - }); -#else - PYOPENCL_UNSUPPORTED(clCreateSubBuffer, "CL 1.0") -#endif -} diff --git a/src/c_wrapper/buffer.h b/src/c_wrapper/buffer.h deleted file mode 100644 index c97a7919..00000000 --- a/src/c_wrapper/buffer.h +++ /dev/null @@ -1,27 +0,0 @@ -#include "memory_object.h" -#include "clhelper.h" - -#ifndef __PYOPENCL_BUFFER_H -#define __PYOPENCL_BUFFER_H - -// {{{ buffer - -class buffer : public memory_object { -public: - PYOPENCL_DEF_CL_CLASS(BUFFER); - PYOPENCL_INLINE - buffer(cl_mem mem, bool retain) - : memory_object(mem, retain) - {} - -#if PYOPENCL_CL_VERSION >= 0x1010 - PYOPENCL_USE_RESULT buffer *get_sub_region(size_t orig, size_t size, - cl_mem_flags flags) const; -#endif -}; - -extern template void print_clobj(std::ostream&, const buffer*); - -// }}} - -#endif diff --git a/src/c_wrapper/clhelper.h b/src/c_wrapper/clhelper.h deleted file mode 100644 index d0aff85c..00000000 --- a/src/c_wrapper/clhelper.h +++ /dev/null @@ -1,254 +0,0 @@ -#include "error.h" -#include "clobj.h" - -#ifndef __PYOPENCL_CLHELPER_H -#define __PYOPENCL_CLHELPER_H - -template -class _CLObjOutArg : public OutArg { - typedef typename CLObj::cl_type CLType; - clobj_t *const m_ret; - CLType m_clobj; - cl_int (CL_API_CALL *m_release)(CLType); - const char *m_name; - std::tuple m_t1; - template - PYOPENCL_INLINE CLObj* - __new_obj(seq) - { - return new CLObj(m_clobj, false, std::get(m_t1)...); - } -public: - PYOPENCL_INLINE - _CLObjOutArg(clobj_t *ret, cl_int (CL_API_CALL *release)(CLType), - const char *name, T... t1) noexcept - : m_ret(ret), m_clobj(nullptr), m_release(release), - m_name(name), m_t1(t1...) - { - } - PYOPENCL_INLINE - _CLObjOutArg(_CLObjOutArg &&other) noexcept - : m_ret(other.m_ret), m_clobj(other.m_clobj), - m_release(other.m_release), m_name(other.m_name) - { - std::swap(m_t1, other.m_t1); - } - PYOPENCL_INLINE typename CLObj::cl_type* - get() - { - return &m_clobj; - } - PYOPENCL_INLINE void - convert() - { - *m_ret = __new_obj(typename gens::type()); - } - PYOPENCL_INLINE void - cleanup(bool converted) - { - if (converted) { - delete *m_ret; - *m_ret = nullptr; - } else { - call_guarded_cleanup(m_release, m_name, m_clobj); - } - } - PYOPENCL_INLINE void - print(std::ostream &stm, bool out=false) const - { - print_arg(stm, m_clobj, out); - } -}; - -template -static PYOPENCL_INLINE _CLObjOutArg -make_cloutarg(clobj_t *ret, cl_int (CL_API_CALL *release)(typename CLObj::cl_type), - const char *name, T... t1) -{ - return _CLObjOutArg(ret, release, name, t1...); -} -#define pyopencl_outarg(type, ret, func, ...) \ - make_cloutarg(ret, func, #func, ##__VA_ARGS__) - -// {{{ GetInfo helpers - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE pyopencl_buf -get_vec_info(cl_int (CL_API_CALL *func)(ArgTypes...), const char *name, - ArgTypes2&&... args) -{ - size_t size = 0; - call_guarded(func, name, args..., 0, nullptr, buf_arg(size)); - pyopencl_buf buf(size / sizeof(T)); - call_guarded(func, name, args..., size_arg(buf), buf_arg(size)); - return buf; -} -#define pyopencl_get_vec_info(type, what, ...) \ - get_vec_info(clGet##what##Info, "clGet" #what "Info", __VA_ARGS__) - -inline generic_info make_generic_info(class_t opaque_class, const char *type, bool free_type, void *value, bool free_value) -{ - generic_info result; - result.opaque_class = opaque_class; - result.type = type; - result.free_type = free_type; - result.value = value; - result.free_value = free_value; - return result; -} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE generic_info -convert_array_info(const char *tname, pyopencl_buf &buf) -{ - return make_generic_info( - CLASS_NONE, - _copy_str(std::string(tname) + "[" + tostring(buf.len()) + "]"), - true, - buf.release(), - true); -} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE generic_info -convert_array_info(const char *tname, pyopencl_buf &&_buf) -{ - pyopencl_buf &buf = _buf; - return convert_array_info(tname, buf); -} - -#define pyopencl_convert_array_info(type, buf) \ - convert_array_info(#type, buf) -#define pyopencl_get_array_info(type, what, ...) \ - pyopencl_convert_array_info(type, pyopencl_get_vec_info(type, what, __VA_ARGS__)) - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE generic_info -convert_opaque_array_info(T &&buf) -{ - return make_generic_info( - CLObj::class_id, - _copy_str(std::string("void*[") + tostring(buf.len()) + "]"), - true, - buf_to_base(std::forward(buf)).release(), - true); -} -#define pyopencl_get_opaque_array_info(cls, what, ...) \ - convert_opaque_array_info( \ - pyopencl_get_vec_info(cls::cl_type, what, __VA_ARGS__)) - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE generic_info -get_opaque_info(cl_int (CL_API_CALL *func)(ArgTypes...), const char *name, - ArgTypes2&&... args) -{ - typename CLObj::cl_type param_value; - call_guarded(func, name, args..., size_arg(param_value), nullptr); - void *value; - if (param_value) { - value = (void*)(new CLObj(param_value, /*retain*/ true)); - } else { - value = nullptr; - } - return make_generic_info(CLObj::class_id, "void *", false, value, true); -} -#define pyopencl_get_opaque_info(clobj, what, ...) \ - get_opaque_info(clGet##what##Info, \ - "clGet" #what "Info", __VA_ARGS__) - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE generic_info -get_str_info(cl_int (CL_API_CALL *func)(ArgTypes...), const char *name, - ArgTypes2&&... args) -{ - size_t size; - call_guarded(func, name, args..., 0, nullptr, buf_arg(size)); - pyopencl_buf param_value(size); - call_guarded(func, name, args..., param_value, buf_arg(size)); - return make_generic_info(CLASS_NONE, "char*", false, (void*)param_value.release(), true); -} -#define pyopencl_get_str_info(what, ...) \ - get_str_info(clGet##what##Info, "clGet" #what "Info", __VA_ARGS__) - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE generic_info -get_int_info(cl_int (CL_API_CALL *func)(ArgTypes...), const char *name, - const char *tpname, ArgTypes2&&... args) -{ - T value; - call_guarded(func, name, args..., size_arg(value), nullptr); - return make_generic_info(CLASS_NONE, tpname, false, cl_memdup(&value), true); -} -#define pyopencl_get_int_info(type, what, ...) \ - get_int_info(clGet##what##Info, "clGet" #what "Info", \ - #type "*", __VA_ARGS__) - -// }}} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE T* -convert_obj(cl_int (CL_API_CALL *clRelease)(CLType), const char *name, CLType cl_obj, - ArgTypes&&... args) -{ - try { - return new T(cl_obj, false, std::forward(args)...); - } catch (...) { - call_guarded_cleanup(clRelease, name, cl_obj); - throw; - } -} -#define pyopencl_convert_obj(type, func, ...) \ - convert_obj(func, #func, __VA_ARGS__) - -// {{{ extension function pointers - -#if PYOPENCL_CL_VERSION >= 0x1020 -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE T -get_ext_fun(cl_platform_id plat, const char *name, const char *err) -{ - T func = (T)clGetExtensionFunctionAddressForPlatform(plat, name); - if (!func) { - throw clerror(name, CL_INVALID_VALUE, err); - } - return func; -} -#define pyopencl_get_ext_fun(plat, name) \ - get_ext_fun(plat, #name, #name " not available") -#else -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE T -get_ext_fun(const char *name, const char *err) -{ - T func = (T)clGetExtensionFunctionAddress(name); - if (!func) { - throw clerror(name, CL_INVALID_VALUE, err); - } - return func; -} -#define pyopencl_get_ext_fun(plat, name) \ - get_ext_fun(#name, #name " not available") -#endif - -// }}} - -static PYOPENCL_INLINE std::ostream& -operator<<(std::ostream &stm, const cl_image_format &fmt) -{ - stm << "channel_order: " << fmt.image_channel_order - << ",\nchannel_data_type: " << fmt.image_channel_data_type; - return stm; -} - -#ifdef CL_DEVICE_TOPOLOGY_AMD -static PYOPENCL_INLINE std::ostream& -operator<<(std::ostream &stm, const cl_device_topology_amd &topol) -{ - stm << "pcie.bus: " << topol.pcie.bus - << ",\npcie.device: " << topol.pcie.device - << ",\npcie.function: " << topol.pcie.function - << ",\npcie.type: " << topol.pcie.type; - return stm; -} -#endif -#endif diff --git a/src/c_wrapper/clinfo_ext.h b/src/c_wrapper/clinfo_ext.h deleted file mode 100644 index 43b7b608..00000000 --- a/src/c_wrapper/clinfo_ext.h +++ /dev/null @@ -1,129 +0,0 @@ -/* Include OpenCL header, and define OpenCL extensions, since what is and is not - * available in the official headers is very system-dependent */ - -#ifndef _EXT_H -#define _EXT_H - -#if (defined(__APPLE__) && !defined(PYOPENCL_APPLE_USE_CL_H)) -#include -#else -#include -#endif - -/* These two defines were introduced in the 1.2 headers - * on 2012-11-30, so earlier versions don't have them - * (e.g. Debian wheezy) - */ - -#ifndef CL_DEVICE_IMAGE_PITCH_ALIGNMENT -#define CL_DEVICE_IMAGE_PITCH_ALIGNMENT 0x104A -#define CL_DEVICE_IMAGE_BASE_ADDRESS_ALIGNMENT 0x104B -#endif - -/* - * Extensions - */ - -/* cl_khr_icd */ -#define CL_PLATFORM_ICD_SUFFIX_KHR 0x0920 -#define CL_PLATFORM_NOT_FOUND_KHR -1001 - - -/* cl_khr_fp64 */ -#define CL_DEVICE_DOUBLE_FP_CONFIG 0x1032 - -/* cl_khr_fp16 */ -#define CL_DEVICE_HALF_FP_CONFIG 0x1033 - -/* cl_khr_terminate_context */ -#define CL_DEVICE_TERMINATE_CAPABILITY_KHR 0x200F - -/* cl_nv_device_attribute_query */ -#define CL_DEVICE_COMPUTE_CAPABILITY_MAJOR_NV 0x4000 -#define CL_DEVICE_COMPUTE_CAPABILITY_MINOR_NV 0x4001 -#define CL_DEVICE_REGISTERS_PER_BLOCK_NV 0x4002 -#define CL_DEVICE_WARP_SIZE_NV 0x4003 -#define CL_DEVICE_GPU_OVERLAP_NV 0x4004 -#define CL_DEVICE_KERNEL_EXEC_TIMEOUT_NV 0x4005 -#define CL_DEVICE_INTEGRATED_MEMORY_NV 0x4006 -#define CL_DEVICE_ATTRIBUTE_ASYNC_ENGINE_COUNT_NV 0x4007 -#define CL_DEVICE_PCI_BUS_ID_NV 0x4008 -#define CL_DEVICE_PCI_SLOT_ID_NV 0x4009 - -/* cl_ext_atomic_counters_{32,64} */ -#define CL_DEVICE_MAX_ATOMIC_COUNTERS_EXT 0x4032 - -/* cl_amd_device_attribute_query */ -#define CL_DEVICE_PROFILING_TIMER_OFFSET_AMD 0x4036 -#define CL_DEVICE_TOPOLOGY_AMD 0x4037 -#define CL_DEVICE_BOARD_NAME_AMD 0x4038 -#define CL_DEVICE_GLOBAL_FREE_MEMORY_AMD 0x4039 -#define CL_DEVICE_SIMD_PER_COMPUTE_UNIT_AMD 0x4040 -#define CL_DEVICE_SIMD_WIDTH_AMD 0x4041 -#define CL_DEVICE_SIMD_INSTRUCTION_WIDTH_AMD 0x4042 -#define CL_DEVICE_WAVEFRONT_WIDTH_AMD 0x4043 -#define CL_DEVICE_GLOBAL_MEM_CHANNELS_AMD 0x4044 -#define CL_DEVICE_GLOBAL_MEM_CHANNEL_BANKS_AMD 0x4045 -#define CL_DEVICE_GLOBAL_MEM_CHANNEL_BANK_WIDTH_AMD 0x4046 -#define CL_DEVICE_LOCAL_MEM_SIZE_PER_COMPUTE_UNIT_AMD 0x4047 -#define CL_DEVICE_LOCAL_MEM_BANKS_AMD 0x4048 -#define CL_DEVICE_THREAD_TRACE_SUPPORTED_AMD 0x4049 -#define CL_DEVICE_GFXIP_MAJOR_AMD 0x404A -#define CL_DEVICE_GFXIP_MINOR_AMD 0x404B -#define CL_DEVICE_AVAILABLE_ASYNC_QUEUES_AMD 0x404C - -#ifndef CL_DEVICE_TOPOLOGY_TYPE_PCIE_AMD -#define CL_DEVICE_TOPOLOGY_TYPE_PCIE_AMD 1 - -typedef union -{ - struct { cl_uint type; cl_uint data[5]; } raw; - struct { cl_uint type; cl_char unused[17]; cl_char bus; cl_char device; cl_char function; } pcie; -} cl_device_topology_amd; -#endif - -/* cl_amd_offline_devices */ -#define CL_CONTEXT_OFFLINE_DEVICES_AMD 0x403F - -/* cl_ext_device_fission */ -#define cl_ext_device_fission 1 - -typedef cl_ulong cl_device_partition_property_ext; - -#define CL_DEVICE_PARTITION_EQUALLY_EXT 0x4050 -#define CL_DEVICE_PARTITION_BY_COUNTS_EXT 0x4051 -#define CL_DEVICE_PARTITION_BY_NAMES_EXT 0x4052 -#define CL_DEVICE_PARTITION_BY_NAMES_INTEL 0x4052 /* cl_intel_device_partition_by_names */ -#define CL_DEVICE_PARTITION_BY_AFFINITY_DOMAIN_EXT 0x4053 - -#define CL_DEVICE_PARENT_DEVICE_EXT 0x4054 -#define CL_DEVICE_PARTITION_TYPES_EXT 0x4055 -#define CL_DEVICE_AFFINITY_DOMAINS_EXT 0x4056 -#define CL_DEVICE_REFERENCE_COUNT_EXT 0x4057 -#define CL_DEVICE_PARTITION_STYLE_EXT 0x4058 - -#define CL_AFFINITY_DOMAIN_L1_CACHE_EXT 0x1 -#define CL_AFFINITY_DOMAIN_L2_CACHE_EXT 0x2 -#define CL_AFFINITY_DOMAIN_L3_CACHE_EXT 0x3 -#define CL_AFFINITY_DOMAIN_L4_CACHE_EXT 0x4 -#define CL_AFFINITY_DOMAIN_NUMA_EXT 0x10 -#define CL_AFFINITY_DOMAIN_NEXT_FISSIONABLE_EXT 0x100 - -/* cl_intel_advanced_motion_estimation */ -#define CL_DEVICE_ME_VERSION_INTEL 0x407E - -/* cl_qcom_ext_host_ptr */ -#define CL_DEVICE_EXT_MEM_PADDING_IN_BYTES_QCOM 0x40A0 -#define CL_DEVICE_PAGE_SIZE_QCOM 0x40A1 - -/* cl_khr_spir */ -#define CL_DEVICE_SPIR_VERSIONS 0x40E0 - -/* cl_altera_device_temperature */ -#define CL_DEVICE_CORE_TEMPERATURE_ALTERA 0x40F3 - -/* cl_intel_simultaneous_sharing */ -#define CL_DEVICE_SIMULTANEOUS_INTEROPS_INTEL 0x4104 -#define CL_DEVICE_NUM_SIMULTANEOUS_INTEROPS_INTEL 0x4105 - -#endif diff --git a/src/c_wrapper/clobj.h b/src/c_wrapper/clobj.h deleted file mode 100644 index 5db08710..00000000 --- a/src/c_wrapper/clobj.h +++ /dev/null @@ -1,149 +0,0 @@ -#include "utils.h" - -#ifndef __PYOPENCL_CLOBJ_H -#define __PYOPENCL_CLOBJ_H - -#define PYOPENCL_DEF_CL_CLASS(name) \ - constexpr static class_t class_id = CLASS_##name; \ - constexpr static const char *class_name = #name; - -struct clbase { -private: - // non-copyable - clbase(const clbase&) = delete; - clbase &operator=(const clbase&) = delete; - bool operator==(clbase const &other) const = delete; - bool operator!=(clbase const &other) const = delete; -public: - clbase() = default; - virtual ~clbase() = default; - virtual intptr_t intptr() const = 0; - virtual generic_info get_info(cl_uint) const = 0; -}; - -template -class clobj : public clbase { -private: - CLType m_obj; -public: - typedef CLType cl_type; - PYOPENCL_INLINE - clobj(CLType obj, bool=false) : m_obj(obj) - {} - PYOPENCL_INLINE const CLType& - data() const - { - return m_obj; - } - intptr_t - intptr() const - { - return (intptr_t)m_obj; - } -}; - -template -void -print_clobj(std::ostream &stm, const CLObj *obj) -{ - stm << CLObj::class_name << "(" << (const void*)obj << ")<" - << (const void*)obj->data() << ">"; -} - -template -class CLArg, - CLObj>::value> > { -private: - CLObj &m_obj; -public: - CLArg(CLObj &obj) : m_obj(obj) - { - } - PYOPENCL_INLINE const typename CLObj::cl_type& - convert() const - { - return m_obj.data(); - } - PYOPENCL_INLINE void - print(std::ostream &stm) - { - print_clobj(stm, &m_obj); - } -}; - -template -class CLArg, - CLObj>::value> > { -private: - CLObj *m_obj; -public: - CLArg(CLObj *obj) : m_obj(obj) - { - } - PYOPENCL_INLINE const typename CLObj::cl_type& - convert() const - { - return m_obj->data(); - } - PYOPENCL_INLINE void - print(std::ostream &stm) - { - print_clobj(stm, m_obj); - } -}; - -template -static PYOPENCL_INLINE CLObj* -clobj_from_int_ptr(intptr_t ptr, bool retain) -{ - return new CLObj(reinterpret_cast(ptr), retain); -} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE pyopencl_buf -buf_from_class(T2 *buf2, size_t len) -{ - pyopencl_buf buf(len); - for (size_t i = 0;i < len;i++) { - buf[i] = static_cast(buf2[i])->data(); - } - return buf; -} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE pyopencl_buf -buf_from_class(T2 &&buf) -{ - return buf_from_class(buf.get(), buf.len()); -} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE pyopencl_buf -buf_to_base(T2 *buf2, size_t len, ArgTypes&&... args) -{ - pyopencl_buf buf(len); - size_t i = 0; - try { - for (;i < len;i++) { - buf[i] = static_cast( - new T((typename T::cl_type)buf2[i], - std::forward(args)...)); - } - } catch (...) { - for (size_t j = 0;j < i;j++) { - delete buf[i]; - } - throw; - } - return buf; -} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE pyopencl_buf -buf_to_base(T2 &&buf2, ArgTypes&&... args) -{ - return buf_to_base(buf2.get(), buf2.len(), - std::forward(args)...); -} - -#endif diff --git a/src/c_wrapper/command_queue.cpp b/src/c_wrapper/command_queue.cpp deleted file mode 100644 index b8ecef1e..00000000 --- a/src/c_wrapper/command_queue.cpp +++ /dev/null @@ -1,132 +0,0 @@ -#include "command_queue.h" -#include "device.h" -#include "context.h" -#include "event.h" -#include "clhelper.h" - -template class clobj; -template void print_arg(std::ostream&, - const cl_command_queue&, bool); -template void print_clobj(std::ostream&, const command_queue*); -template void print_buf( - std::ostream&, const cl_command_queue*, size_t, ArgType, bool, bool); - -command_queue::~command_queue() -{ - pyopencl_call_guarded_cleanup(clReleaseCommandQueue, PYOPENCL_CL_CASTABLE_THIS); -} - -generic_info -command_queue::get_info(cl_uint param_name) const -{ - switch ((cl_command_queue_info)param_name) { - case CL_QUEUE_CONTEXT: - return pyopencl_get_opaque_info(context, CommandQueue, - PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_QUEUE_DEVICE: - return pyopencl_get_opaque_info(device, CommandQueue, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_QUEUE_REFERENCE_COUNT: - return pyopencl_get_int_info(cl_uint, CommandQueue, - PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_QUEUE_PROPERTIES: - return pyopencl_get_int_info(cl_command_queue_properties, - CommandQueue, PYOPENCL_CL_CASTABLE_THIS, param_name); - default: - throw clerror("CommandQueue.get_info", CL_INVALID_VALUE); - } -} - -// c wrapper - -// Command Queue -error* -create_command_queue(clobj_t *queue, clobj_t _ctx, - clobj_t _dev, cl_command_queue_properties props) -{ - auto ctx = static_cast(_ctx); - auto py_dev = static_cast(_dev); - return c_handle_error([&] { - cl_device_id dev; - if (py_dev) { - dev = py_dev->data(); - } else { - auto devs = pyopencl_get_vec_info(cl_device_id, Context, - ctx, CL_CONTEXT_DEVICES); - if (devs.len() == 0) { - throw clerror("CommandQueue", CL_INVALID_VALUE, - "context doesn't have any devices? -- " - "don't know which one to default to"); - } - dev = devs[0]; - } - cl_command_queue cl_queue = - pyopencl_call_guarded(clCreateCommandQueue, ctx, dev, props); - *queue = new command_queue(cl_queue, false); - }); -} - -error* -command_queue__finish(clobj_t queue) -{ - return c_handle_error([&] { - pyopencl_call_guarded(clFinish, static_cast(queue)); - }); -} - -error* -command_queue__flush(clobj_t queue) -{ - return c_handle_error([&] { - pyopencl_call_guarded(clFlush, static_cast(queue)); - }); -} - -error* -enqueue_marker_with_wait_list(clobj_t *evt, clobj_t _queue, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - auto queue = static_cast(_queue); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - return c_handle_error([&] { - pyopencl_call_guarded(clEnqueueMarkerWithWaitList, queue, - wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clEnqueueMarkerWithWaitList, "CL 1.2") -#endif -} - -error* -enqueue_barrier_with_wait_list(clobj_t *evt, clobj_t _queue, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - auto queue = static_cast(_queue); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - return c_handle_error([&] { - pyopencl_call_guarded(clEnqueueBarrierWithWaitList, queue, - wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clEnqueueBarrierWithWaitList, "CL 1.2") -#endif -} - -error* -enqueue_marker(clobj_t *evt, clobj_t _queue) -{ - auto queue = static_cast(_queue); - return c_handle_error([&] { - pyopencl_call_guarded(clEnqueueMarker, queue, event_out(evt)); - }); -} - -error* -enqueue_barrier(clobj_t _queue) -{ - auto queue = static_cast(_queue); - return c_handle_error([&] { - pyopencl_call_guarded(clEnqueueBarrier, queue); - }); -} diff --git a/src/c_wrapper/command_queue.h b/src/c_wrapper/command_queue.h deleted file mode 100644 index 3a7c0171..00000000 --- a/src/c_wrapper/command_queue.h +++ /dev/null @@ -1,64 +0,0 @@ -#include "error.h" - -#ifndef __PYOPENCL_COMMAND_QUEUE_H -#define __PYOPENCL_COMMAND_QUEUE_H - -// {{{ command_queue - -extern template class clobj; -extern template void print_arg( - std::ostream&, const cl_command_queue&, bool); -extern template void print_buf( - std::ostream&, const cl_command_queue*, size_t, ArgType, bool, bool); - -class command_queue : public clobj { -public: - PYOPENCL_DEF_CL_CLASS(COMMAND_QUEUE); - PYOPENCL_INLINE - command_queue(cl_command_queue q, bool retain) - : clobj(q) - { - if (retain) { - pyopencl_call_guarded(clRetainCommandQueue, PYOPENCL_CL_CASTABLE_THIS); - } - } - PYOPENCL_INLINE - command_queue(const command_queue &queue) - : command_queue(queue.data(), true) - {} - ~command_queue(); - - generic_info get_info(cl_uint param_name) const; - -#if 0 - - PYOPENCL_USE_RESULT std::unique_ptr - get_context() const - { - cl_context param_value; - pyopencl_call_guarded(clGetCommandQueueInfo, this, CL_QUEUE_CONTEXT, - size_arg(param_value), nullptr); - return std::unique_ptr( - new context(param_value, /*retain*/ true)); - } - -#if PYOPENCL_CL_VERSION < 0x1010 - cl_command_queue_properties - set_property(cl_command_queue_properties prop, bool enable) const - { - cl_command_queue_properties old_prop; - pyopencl_call_guarded(clSetCommandQueueProperty, this, prop, - enable, buf_arg(old_prop)); - return old_prop; - } -#endif - -#endif -}; - -extern template void print_clobj(std::ostream&, - const command_queue*); - -// }}} - -#endif diff --git a/src/c_wrapper/context.cpp b/src/c_wrapper/context.cpp deleted file mode 100644 index 0fe48554..00000000 --- a/src/c_wrapper/context.cpp +++ /dev/null @@ -1,153 +0,0 @@ -#include "context.h" -#include "device.h" -#include "platform.h" -#include "clhelper.h" - -template class clobj; -template void print_arg(std::ostream&, const cl_context&, bool); -template void print_clobj(std::ostream&, const context*); -template void print_buf(std::ostream&, const cl_context*, - size_t, ArgType, bool, bool); - -void -context::get_version(cl_context ctx, int *major, int *minor) -{ - cl_device_id s_buff[16]; - size_t size; - pyopencl_buf d_buff(0); - cl_device_id *devs = s_buff; - pyopencl_call_guarded(clGetContextInfo, ctx, CL_CONTEXT_DEVICES, - 0, nullptr, buf_arg(size)); - if (PYOPENCL_UNLIKELY(!size)) { - throw clerror("Context.get_version", CL_INVALID_VALUE, - "Cannot get devices from context."); - } - if (PYOPENCL_UNLIKELY(size > sizeof(s_buff))) { - d_buff.resize(size / sizeof(cl_device_id)); - devs = d_buff.get(); - } - pyopencl_call_guarded(clGetContextInfo, ctx, CL_CONTEXT_DEVICES, - size_arg(devs, size), buf_arg(size)); - device::get_version(devs[0], major, minor); -} - -context::~context() -{ - pyopencl_call_guarded_cleanup(clReleaseContext, PYOPENCL_CL_CASTABLE_THIS); -} - -generic_info -context::get_info(cl_uint param_name) const -{ - switch ((cl_context_info)param_name) { - case CL_CONTEXT_REFERENCE_COUNT: - return pyopencl_get_int_info(cl_uint, Context, - PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_CONTEXT_DEVICES: - return pyopencl_get_opaque_array_info(device, Context, - PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_CONTEXT_PROPERTIES: { - auto result = pyopencl_get_vec_info( - cl_context_properties, Context, PYOPENCL_CL_CASTABLE_THIS, param_name); - pyopencl_buf py_result(result.len() / 2); - size_t i = 0; - for (;i < py_result.len();i++) { - cl_context_properties key = result[i * 2]; - if (key == 0) - break; - cl_context_properties value = result[i * 2 + 1]; - switch (key) { - case CL_CONTEXT_PLATFORM: - py_result[i] = make_generic_info( - CLASS_PLATFORM, - "void *", false, - new platform(reinterpret_cast(value)), true); - break; - -#if defined(PYOPENCL_GL_SHARING_VERSION) && (PYOPENCL_GL_SHARING_VERSION >= 1) -#if defined(__APPLE__) && defined(HAVE_GL) && !defined(PYOPENCL_APPLE_USE_CL_H) - case CL_CONTEXT_PROPERTY_USE_CGL_SHAREGROUP_APPLE: -#else - case CL_GL_CONTEXT_KHR: - case CL_EGL_DISPLAY_KHR: - case CL_GLX_DISPLAY_KHR: - case CL_WGL_HDC_KHR: - case CL_CGL_SHAREGROUP_KHR: -#endif - py_result[i] = make_generic_info( - CLASS_NONE, - "intptr_t *", false, - (void*)value, - // we do not own this object - false); - break; -#endif - default: - throw clerror("Context.get_info", CL_INVALID_VALUE, - "unknown context_property key encountered"); - } - } - py_result.resize(i); - return pyopencl_convert_array_info(generic_info, py_result); - } - -#if PYOPENCL_CL_VERSION >= 0x1010 - case CL_CONTEXT_NUM_DEVICES: - return pyopencl_get_int_info(cl_uint, Context, - PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif - - default: - throw clerror("Context.get_info", CL_INVALID_VALUE); - } -} - -// c wrapper - -// Context -error* -create_context(clobj_t *_ctx, const cl_context_properties *props, - cl_uint num_devices, const clobj_t *_devices) -{ - // TODO debug print properties - return c_handle_error([&] { - const auto devices = buf_from_class(_devices, num_devices); - *_ctx = new context( - pyopencl_call_guarded( - clCreateContext, - const_cast(props), - devices, nullptr, nullptr), false); - }); -} - -// Context -error* -create_context_from_type(clobj_t *_ctx, const cl_context_properties *props, - cl_device_type dev_type) -{ - // TODO debug print properties - return c_handle_error([&] { - *_ctx = new context( - pyopencl_call_guarded( - clCreateContextFromType, - const_cast(props), - dev_type, nullptr, nullptr), false); - }); -} - -error* -context__get_supported_image_formats(clobj_t _ctx, cl_mem_flags flags, - cl_mem_object_type image_type, - generic_info *out) -{ - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - cl_uint num; - pyopencl_call_guarded(clGetSupportedImageFormats, ctx, flags, - image_type, 0, nullptr, buf_arg(num)); - pyopencl_buf formats(num); - pyopencl_call_guarded(clGetSupportedImageFormats, ctx, flags, - image_type, formats, buf_arg(num)); - *out = pyopencl_convert_array_info(cl_image_format, formats); - }); -} diff --git a/src/c_wrapper/context.h b/src/c_wrapper/context.h deleted file mode 100644 index 1691035d..00000000 --- a/src/c_wrapper/context.h +++ /dev/null @@ -1,34 +0,0 @@ -#include "error.h" - -#ifndef __PYOPENCL_CONTEXT_H -#define __PYOPENCL_CONTEXT_H - -// {{{ context - -extern template class clobj; -extern template void print_arg(std::ostream&, - const cl_context&, bool); -extern template void print_buf(std::ostream&, const cl_context*, - size_t, ArgType, bool, bool); - -class context : public clobj { -public: - PYOPENCL_DEF_CL_CLASS(CONTEXT); - static void get_version(cl_context ctx, int *major, int *minor); - PYOPENCL_INLINE - context(cl_context ctx, bool retain) - : clobj(ctx) - { - if (retain) { - pyopencl_call_guarded(clRetainContext, PYOPENCL_CL_CASTABLE_THIS); - } - } - ~context(); - generic_info get_info(cl_uint param_name) const; -}; - -extern template void print_clobj(std::ostream&, const context*); - -// }}} - -#endif diff --git a/src/c_wrapper/debug.cpp b/src/c_wrapper/debug.cpp deleted file mode 100644 index a118b468..00000000 --- a/src/c_wrapper/debug.cpp +++ /dev/null @@ -1,84 +0,0 @@ -#include "debug.h" -#include -#include -#include -#include - -std::mutex dbg_lock; - -void -dbg_print_str(std::ostream &stm, const char *str, size_t len) -{ - stm << '"'; - for (size_t i = 0;i < len;i++) { - char escaped = 0; -#define escape_char(in, out) \ - case in: \ - escaped = out; \ - break - switch (str[i]) { - escape_char('\'', '\''); - escape_char('\"', '\"'); - escape_char('\?', '\?'); - escape_char('\\', '\\'); - escape_char('\0', '0'); - escape_char('\a', 'a'); - escape_char('\b', 'b'); - escape_char('\f', 'f'); - escape_char('\r', 'r'); - escape_char('\v', 'v'); - default: - break; - } - if (escaped) { - stm << '\\' << escaped; - } else { - stm << str[i]; - } - } - stm << '"'; -} - -void -dbg_print_bytes(std::ostream &stm, const unsigned char *bytes, size_t len) -{ - stm << '"'; - for (size_t i = 0;i < len;i++) { - stm << "\\x" << std::hex << std::setfill('0') - << std::setw(2) << bytes[i]; - } - stm << std::dec << '"'; -} - -static PYOPENCL_INLINE bool -_get_debug_env() -{ - const char *env = getenv("PYOPENCL_DEBUG"); - const bool default_debug = DEFAULT_DEBUG; - if (!env) { - return default_debug; - } - if (strcasecmp(env, "0") == 0 || strcasecmp(env, "f") == 0 || - strcasecmp(env, "false") == 0 || strcasecmp(env, "off") == 0) { - return false; - } - if (strcasecmp(env, "1") == 0 || strcasecmp(env, "t") == 0 || - strcasecmp(env, "true") == 0 || strcasecmp(env, "on") == 0) { - return true; - } - return default_debug; -} - -bool debug_enabled = _get_debug_env(); - -int -get_debug() -{ - return (int) debug_enabled; -} - -void -set_debug(int debug) -{ - debug_enabled = (bool)debug; -} diff --git a/src/c_wrapper/debug.h b/src/c_wrapper/debug.h deleted file mode 100644 index f0700030..00000000 --- a/src/c_wrapper/debug.h +++ /dev/null @@ -1,33 +0,0 @@ -#include "wrap_cl.h" -#include "function.h" -#include -#include - -#ifdef __MINGW32__ -#include "mingw-std-threads/mingw.mutex.h" -#include "mingw-std-threads/mingw.thread.h" -#endif - -#ifndef __PYOPENCL_DEBUG_H -#define __PYOPENCL_DEBUG_H - -extern bool debug_enabled; -#ifdef PYOPENCL_TRACE -#define DEFAULT_DEBUG true -#else -#define DEFAULT_DEBUG false -#endif - -#define DEBUG_ON (PYOPENCL_EXPECT(debug_enabled, DEFAULT_DEBUG)) - -extern std::mutex dbg_lock; - -void dbg_print_str(std::ostream&, const char*, size_t); -static PYOPENCL_INLINE void -dbg_print_str(std::ostream &stm, const char *str) -{ - return dbg_print_str(stm, str, strlen(str)); -} -void dbg_print_bytes(std::ostream &stm, const unsigned char *bytes, size_t len); - -#endif diff --git a/src/c_wrapper/device.cpp b/src/c_wrapper/device.cpp deleted file mode 100644 index 16edaf34..00000000 --- a/src/c_wrapper/device.cpp +++ /dev/null @@ -1,375 +0,0 @@ -#include "device.h" -#include "platform.h" - -template class clobj; -template void print_arg(std::ostream&, - const cl_device_id&, bool); -template void print_clobj(std::ostream&, const device*); -template void print_buf(std::ostream&, const cl_device_id*, - size_t, ArgType, bool, bool); - -void -device::get_version(cl_device_id dev, int *major, int *minor) -{ - cl_platform_id plat; - pyopencl_call_guarded(clGetDeviceInfo, dev, CL_DEVICE_PLATFORM, - size_arg(plat), nullptr); - platform::get_version(plat, major, minor); -} - -device::~device() -{ - if (false) { - } -#if PYOPENCL_CL_VERSION >= 0x1020 - else if (m_ref_type == REF_CL_1_2) { - pyopencl_call_guarded_cleanup(clReleaseDevice, PYOPENCL_CL_CASTABLE_THIS); - } -#endif -} - -#ifdef CL_DEVICE_TOPOLOGY_AMD -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE generic_info -get_device_topology_amd(ArgTypes&&... args) -{ - const char * tpname = "cl_device_topology_amd*"; - cl_device_topology_amd value; - const char * fname = "clGetDeviceInfo"; - call_guarded(clGetDeviceInfo, fname, args..., size_arg(value), nullptr); - return make_generic_info(CLASS_NONE, tpname, false, cl_memdup(&value), true); -} - -#define pyopencl_get_device_topology_amd(...) get_device_topology_amd(__VA_ARGS__) - -#endif - -generic_info -device::get_info(cl_uint param_name) const -{ -#define DEV_GET_INT_INF(TYPE) \ - pyopencl_get_int_info(TYPE, Device, PYOPENCL_CL_CASTABLE_THIS, param_name) - - switch ((cl_device_info)param_name) { - case CL_DEVICE_TYPE: - return DEV_GET_INT_INF(cl_device_type); - case CL_DEVICE_MAX_WORK_GROUP_SIZE: - return DEV_GET_INT_INF(size_t); - case CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS: - case CL_DEVICE_MAX_COMPUTE_UNITS: - case CL_DEVICE_VENDOR_ID: - return DEV_GET_INT_INF(cl_uint); - - case CL_DEVICE_MAX_WORK_ITEM_SIZES: - return pyopencl_get_array_info(size_t, Device, PYOPENCL_CL_CASTABLE_THIS, param_name); - - case CL_DEVICE_PREFERRED_VECTOR_WIDTH_CHAR: - case CL_DEVICE_PREFERRED_VECTOR_WIDTH_SHORT: - case CL_DEVICE_PREFERRED_VECTOR_WIDTH_INT: - case CL_DEVICE_PREFERRED_VECTOR_WIDTH_LONG: - case CL_DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT: - case CL_DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE: - - case CL_DEVICE_MAX_CLOCK_FREQUENCY: - case CL_DEVICE_ADDRESS_BITS: - case CL_DEVICE_MAX_READ_IMAGE_ARGS: - case CL_DEVICE_MAX_WRITE_IMAGE_ARGS: - case CL_DEVICE_MAX_SAMPLERS: - case CL_DEVICE_MEM_BASE_ADDR_ALIGN: - case CL_DEVICE_MIN_DATA_TYPE_ALIGN_SIZE: - return DEV_GET_INT_INF(cl_uint); - - case CL_DEVICE_MAX_MEM_ALLOC_SIZE: - return DEV_GET_INT_INF(cl_ulong); - - case CL_DEVICE_IMAGE2D_MAX_WIDTH: - case CL_DEVICE_IMAGE2D_MAX_HEIGHT: - case CL_DEVICE_IMAGE3D_MAX_WIDTH: - case CL_DEVICE_IMAGE3D_MAX_HEIGHT: - case CL_DEVICE_IMAGE3D_MAX_DEPTH: - case CL_DEVICE_MAX_PARAMETER_SIZE: - return DEV_GET_INT_INF(size_t); - - case CL_DEVICE_IMAGE_SUPPORT: - return DEV_GET_INT_INF(cl_bool); -#ifdef CL_DEVICE_DOUBLE_FP_CONFIG - case CL_DEVICE_DOUBLE_FP_CONFIG: -#endif -#ifdef CL_DEVICE_HALF_FP_CONFIG - case CL_DEVICE_HALF_FP_CONFIG: -#endif - case CL_DEVICE_SINGLE_FP_CONFIG: - return DEV_GET_INT_INF(cl_device_fp_config); - - case CL_DEVICE_GLOBAL_MEM_CACHE_TYPE: - return DEV_GET_INT_INF(cl_device_mem_cache_type); - case CL_DEVICE_GLOBAL_MEM_CACHELINE_SIZE: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_GLOBAL_MEM_CACHE_SIZE: - case CL_DEVICE_GLOBAL_MEM_SIZE: - case CL_DEVICE_LOCAL_MEM_SIZE: - case CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE: - return DEV_GET_INT_INF(cl_ulong); - - case CL_DEVICE_MAX_CONSTANT_ARGS: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_LOCAL_MEM_TYPE: - return DEV_GET_INT_INF(cl_device_local_mem_type); - case CL_DEVICE_PROFILING_TIMER_RESOLUTION: - return DEV_GET_INT_INF(size_t); - case CL_DEVICE_ENDIAN_LITTLE: - case CL_DEVICE_AVAILABLE: - case CL_DEVICE_COMPILER_AVAILABLE: - case CL_DEVICE_ERROR_CORRECTION_SUPPORT: - return DEV_GET_INT_INF(cl_bool); - case CL_DEVICE_EXECUTION_CAPABILITIES: - return DEV_GET_INT_INF(cl_device_exec_capabilities); - case CL_DEVICE_QUEUE_PROPERTIES: - // same as CL_DEVICE_QUEUE_ON_HOST_PROPERTIES in 2.0 - return DEV_GET_INT_INF(cl_command_queue_properties); - - case CL_DEVICE_NAME: - case CL_DEVICE_VENDOR: - case CL_DRIVER_VERSION: - case CL_DEVICE_PROFILE: - case CL_DEVICE_VERSION: - case CL_DEVICE_EXTENSIONS: - return pyopencl_get_str_info(Device, PYOPENCL_CL_CASTABLE_THIS, param_name); - - case CL_DEVICE_PLATFORM: - return pyopencl_get_opaque_info(platform, Device, PYOPENCL_CL_CASTABLE_THIS, param_name); -#if PYOPENCL_CL_VERSION >= 0x1010 - case CL_DEVICE_PREFERRED_VECTOR_WIDTH_HALF: - case CL_DEVICE_NATIVE_VECTOR_WIDTH_CHAR: - case CL_DEVICE_NATIVE_VECTOR_WIDTH_SHORT: - case CL_DEVICE_NATIVE_VECTOR_WIDTH_INT: - case CL_DEVICE_NATIVE_VECTOR_WIDTH_LONG: - case CL_DEVICE_NATIVE_VECTOR_WIDTH_FLOAT: - case CL_DEVICE_NATIVE_VECTOR_WIDTH_DOUBLE: - case CL_DEVICE_NATIVE_VECTOR_WIDTH_HALF: - return DEV_GET_INT_INF(cl_uint); - - case CL_DEVICE_HOST_UNIFIED_MEMORY: // deprecated in 2.0 - return DEV_GET_INT_INF(cl_bool); - case CL_DEVICE_OPENCL_C_VERSION: - return pyopencl_get_str_info(Device, PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif -#ifdef CL_DEVICE_COMPUTE_CAPABILITY_MAJOR_NV - case CL_DEVICE_COMPUTE_CAPABILITY_MAJOR_NV: - case CL_DEVICE_COMPUTE_CAPABILITY_MINOR_NV: - case CL_DEVICE_REGISTERS_PER_BLOCK_NV: - case CL_DEVICE_WARP_SIZE_NV: -#ifdef CL_DEVICE_ATTRIBUTE_ASYNC_ENGINE_COUNT_NV - case CL_DEVICE_ATTRIBUTE_ASYNC_ENGINE_COUNT_NV: -#endif -#ifdef CL_DEVICE_PCI_BUS_ID_NV - case CL_DEVICE_PCI_BUS_ID_NV: -#endif -#ifdef CL_DEVICE_PCI_SLOT_ID_NV - case CL_DEVICE_PCI_SLOT_ID_NV: -#endif - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_GPU_OVERLAP_NV: - case CL_DEVICE_KERNEL_EXEC_TIMEOUT_NV: - case CL_DEVICE_INTEGRATED_MEMORY_NV: - return DEV_GET_INT_INF(cl_bool); -#endif -#if PYOPENCL_CL_VERSION >= 0x1020 - case CL_DEVICE_LINKER_AVAILABLE: - return DEV_GET_INT_INF(cl_bool); - case CL_DEVICE_BUILT_IN_KERNELS: - return pyopencl_get_str_info(Device, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_DEVICE_IMAGE_MAX_BUFFER_SIZE: - case CL_DEVICE_IMAGE_MAX_ARRAY_SIZE: - return DEV_GET_INT_INF(size_t); - case CL_DEVICE_PARENT_DEVICE: - return pyopencl_get_opaque_info(device, Device, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_DEVICE_PARTITION_MAX_SUB_DEVICES: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_PARTITION_TYPE: - case CL_DEVICE_PARTITION_PROPERTIES: - return pyopencl_get_array_info(cl_device_partition_property, - Device, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_DEVICE_PARTITION_AFFINITY_DOMAIN: - return pyopencl_get_array_info(cl_device_affinity_domain, - Device, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_DEVICE_REFERENCE_COUNT: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_PREFERRED_INTEROP_USER_SYNC: - case CL_DEVICE_PRINTF_BUFFER_SIZE: - return DEV_GET_INT_INF(cl_bool); -#endif -#ifdef cl_khr_image2d_from_buffer - case CL_DEVICE_IMAGE_PITCH_ALIGNMENT: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_IMAGE_BASE_ADDRESS_ALIGNMENT: - return DEV_GET_INT_INF(cl_uint); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - case CL_DEVICE_MAX_READ_WRITE_IMAGE_ARGS: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_MAX_GLOBAL_VARIABLE_SIZE: - return DEV_GET_INT_INF(size_t); - case CL_DEVICE_QUEUE_ON_DEVICE_PROPERTIES: - return DEV_GET_INT_INF(cl_command_queue_properties); - case CL_DEVICE_QUEUE_ON_DEVICE_PREFERRED_SIZE: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_QUEUE_ON_DEVICE_MAX_SIZE: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_MAX_ON_DEVICE_QUEUES: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_MAX_ON_DEVICE_EVENTS: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_SVM_CAPABILITIES: - return DEV_GET_INT_INF(cl_device_svm_capabilities); - case CL_DEVICE_GLOBAL_VARIABLE_PREFERRED_TOTAL_SIZE: - return DEV_GET_INT_INF(size_t); - case CL_DEVICE_MAX_PIPE_ARGS: - case CL_DEVICE_PIPE_MAX_ACTIVE_RESERVATIONS: - case CL_DEVICE_PIPE_MAX_PACKET_SIZE: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_PREFERRED_PLATFORM_ATOMIC_ALIGNMENT: - case CL_DEVICE_PREFERRED_GLOBAL_ATOMIC_ALIGNMENT: - case CL_DEVICE_PREFERRED_LOCAL_ATOMIC_ALIGNMENT: - return DEV_GET_INT_INF(cl_uint); -#endif -#if PYOPENCL_CL_VERSION >= 0x2010 - case CL_DEVICE_IL_VERSION: - return pyopencl_get_str_info(Device, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_DEVICE_MAX_NUM_SUB_GROUPS: - return DEV_GET_INT_INF(cl_uint); - case CL_DEVICE_SUB_GROUP_INDEPENDENT_FORWARD_PROGRESS: - return DEV_GET_INT_INF(cl_bool); -#endif - - - // {{{ AMD dev attrs - // - // types of AMD dev attrs divined from - // https://www.khronos.org/registry/cl/api/1.2/cl.hpp -#ifdef CL_DEVICE_PROFILING_TIMER_OFFSET_AMD - case CL_DEVICE_PROFILING_TIMER_OFFSET_AMD: - return DEV_GET_INT_INF(cl_ulong); -#endif -#ifdef CL_DEVICE_TOPOLOGY_AMD - case CL_DEVICE_TOPOLOGY_AMD: - return pyopencl_get_device_topology_amd(PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif -#ifdef CL_DEVICE_THREAD_TRACE_SUPPORTED_AMD - case CL_DEVICE_THREAD_TRACE_SUPPORTED_AMD: - return DEV_GET_INT_INF(cl_bool); -#endif -#ifdef CL_DEVICE_BOARD_NAME_AMD - case CL_DEVICE_BOARD_NAME_AMD: ; - return pyopencl_get_str_info(Device, PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif -#ifdef CL_DEVICE_GLOBAL_FREE_MEMORY_AMD - case CL_DEVICE_GLOBAL_FREE_MEMORY_AMD: - return pyopencl_get_array_info(size_t, Device, - PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif -#ifdef CL_DEVICE_SIMD_PER_COMPUTE_UNIT_AMD - case CL_DEVICE_SIMD_PER_COMPUTE_UNIT_AMD: -#endif -#ifdef CL_DEVICE_SIMD_WIDTH_AMD - case CL_DEVICE_SIMD_WIDTH_AMD: -#endif -#ifdef CL_DEVICE_SIMD_INSTRUCTION_WIDTH_AMD - case CL_DEVICE_SIMD_INSTRUCTION_WIDTH_AMD: -#endif -#ifdef CL_DEVICE_WAVEFRONT_WIDTH_AMD - case CL_DEVICE_WAVEFRONT_WIDTH_AMD: -#endif -#ifdef CL_DEVICE_GLOBAL_MEM_CHANNELS_AMD - case CL_DEVICE_GLOBAL_MEM_CHANNELS_AMD: -#endif -#ifdef CL_DEVICE_GLOBAL_MEM_CHANNEL_BANKS_AMD - case CL_DEVICE_GLOBAL_MEM_CHANNEL_BANKS_AMD: -#endif -#ifdef CL_DEVICE_GLOBAL_MEM_CHANNEL_BANK_WIDTH_AMD - case CL_DEVICE_GLOBAL_MEM_CHANNEL_BANK_WIDTH_AMD: -#endif -#ifdef CL_DEVICE_LOCAL_MEM_SIZE_PER_COMPUTE_UNIT_AMD - case CL_DEVICE_LOCAL_MEM_SIZE_PER_COMPUTE_UNIT_AMD: -#endif -#ifdef CL_DEVICE_LOCAL_MEM_BANKS_AMD - case CL_DEVICE_LOCAL_MEM_BANKS_AMD: -#endif -#ifdef CL_DEVICE_MAX_ATOMIC_COUNTERS_EXT - case CL_DEVICE_MAX_ATOMIC_COUNTERS_EXT: -#endif -#ifdef CL_DEVICE_GFXIP_MAJOR_AMD - case CL_DEVICE_GFXIP_MAJOR_AMD: -#endif -#ifdef CL_DEVICE_GFXIP_MINOR_AMD - case CL_DEVICE_GFXIP_MINOR_AMD: -#endif -#ifdef CL_DEVICE_AVAILABLE_ASYNC_QUEUES_AMD - case CL_DEVICE_AVAILABLE_ASYNC_QUEUES_AMD: -#endif - return DEV_GET_INT_INF(cl_uint); - // }}} -#ifdef CL_DEVICE_ME_VERSION_INTEL - case CL_DEVICE_ME_VERSION_INTEL: -#endif -#ifdef CL_DEVICE_EXT_MEM_PADDING_IN_BYTES_QCOM - case CL_DEVICE_EXT_MEM_PADDING_IN_BYTES_QCOM: -#endif -#ifdef CL_DEVICE_PAGE_SIZE_QCOM - case CL_DEVICE_PAGE_SIZE_QCOM: -#endif -#ifdef CL_DEVICE_NUM_SIMULTANEOUS_INTEROPS_INTEL - case CL_DEVICE_NUM_SIMULTANEOUS_INTEROPS_INTEL: -#endif - return DEV_GET_INT_INF(cl_uint); -#ifdef CL_DEVICE_SIMULTANEOUS_INTEROPS_INTEL - case CL_DEVICE_SIMULTANEOUS_INTEROPS_INTEL: - return pyopencl_get_array_info(cl_uint, Device, PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif -#ifdef CL_DEVICE_SPIR_VERSIONS - case CL_DEVICE_SPIR_VERSIONS: - return pyopencl_get_str_info(Device, PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif -#ifdef CL_DEVICE_CORE_TEMPERATURE_ALTERA - case CL_DEVICE_CORE_TEMPERATURE_ALTERA: - return DEV_GET_INT_INF(cl_int); -#endif - - default: - throw clerror("Device.get_info", CL_INVALID_VALUE); - } -} - -#if PYOPENCL_CL_VERSION >= 0x1020 -PYOPENCL_USE_RESULT pyopencl_buf -device::create_sub_devices(const cl_device_partition_property *props) -{ - // TODO debug print props - cl_uint num_devices; - pyopencl_call_guarded(clCreateSubDevices, PYOPENCL_CL_CASTABLE_THIS, props, 0, nullptr, - buf_arg(num_devices)); - pyopencl_buf devices(num_devices); - pyopencl_call_guarded(clCreateSubDevices, PYOPENCL_CL_CASTABLE_THIS, props, devices, - buf_arg(num_devices)); - return buf_to_base(devices, true, device::REF_CL_1_2); -} -#endif - -// c wrapper - -error* -device__create_sub_devices(clobj_t _dev, clobj_t **_devs, - uint32_t *num_devices, - const cl_device_partition_property *props) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - auto dev = static_cast(_dev); - return c_handle_error([&] { - auto devs = dev->create_sub_devices(props); - *num_devices = (uint32_t)devs.len(); - *_devs = devs.release(); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clCreateSubDevices, "CL 1.2") -#endif -} diff --git a/src/c_wrapper/device.h b/src/c_wrapper/device.h deleted file mode 100644 index a14a9468..00000000 --- a/src/c_wrapper/device.h +++ /dev/null @@ -1,61 +0,0 @@ -#include "clhelper.h" - -#ifndef __PYOPENCL_DEVICE_H -#define __PYOPENCL_DEVICE_H - -// {{{ device - -extern template class clobj; -extern template void print_arg(std::ostream&, - const cl_device_id&, bool); -extern template void print_buf(std::ostream&, const cl_device_id*, - size_t, ArgType, bool, bool); - -class device : public clobj { -public: - PYOPENCL_DEF_CL_CLASS(DEVICE); - enum reference_type_t { - REF_NOT_OWNABLE, - REF_CL_1_2, - }; - -private: - reference_type_t m_ref_type; - -public: - static void get_version(cl_device_id dev, int *major, int *minor); - device(cl_device_id did, bool retain=false, - reference_type_t ref_type=REF_NOT_OWNABLE) - : clobj(did), m_ref_type(ref_type) - { - if (retain && ref_type != REF_NOT_OWNABLE) { - if (false) { - } -#if PYOPENCL_CL_VERSION >= 0x1020 - else if (ref_type == REF_CL_1_2) { - pyopencl_call_guarded(clRetainDevice, PYOPENCL_CL_CASTABLE_THIS); - } -#endif - - else { - throw clerror("Device", CL_INVALID_VALUE, - "cannot own references to devices when device " - "fission or CL 1.2 is not available"); - } - } - } - - ~device(); - - generic_info get_info(cl_uint param_name) const; -#if PYOPENCL_CL_VERSION >= 0x1020 - PYOPENCL_USE_RESULT pyopencl_buf - create_sub_devices(const cl_device_partition_property *props); -#endif -}; - -extern template void print_clobj(std::ostream&, const device*); - -// }}} - -#endif diff --git a/src/c_wrapper/error.h b/src/c_wrapper/error.h deleted file mode 100644 index 30e985f9..00000000 --- a/src/c_wrapper/error.h +++ /dev/null @@ -1,337 +0,0 @@ -#include "wrap_cl.h" -#include "pyhelper.h" -#include "clobj.h" - -#include -#include -#include -#include -#include -#include - -#ifndef __PYOPENCL_ERROR_H -#define __PYOPENCL_ERROR_H - -// {{{ error - -// See https://github.com/inducer/pyopencl/pull/83 -#if GCC_VERSION > 50200 -#define PYOPENCL_CL_CASTABLE_THIS this -#else -#define PYOPENCL_CL_CASTABLE_THIS data() -#endif - -// discouraged, assumes 'version linearity', use PYOPENCL_UNSUPPORTED_BEFORE -#define PYOPENCL_UNSUPPORTED(ROUTINE, VERSION) \ - auto err = (error*)malloc(sizeof(error)); \ - err->routine = strdup(#ROUTINE); \ - err->msg = strdup("unsupported in " VERSION); \ - err->code = CL_INVALID_VALUE; \ - err->other = 0; \ - return err; - -#define PYOPENCL_UNSUPPORTED_BEFORE(ROUTINE, VERSION) \ - auto err = (error*)malloc(sizeof(error)); \ - err->routine = strdup(#ROUTINE); \ - err->msg = strdup("unsupported before " VERSION); \ - err->code = CL_INVALID_VALUE; \ - err->other = 0; \ - return err; - -class clerror : public std::runtime_error { -private: - const char *m_routine; - cl_int m_code; - -public: - clerror(const char *rout, cl_int c, const char *msg="") - : std::runtime_error(msg), m_routine(rout), m_code(c) - { - if (DEBUG_ON) { - std::lock_guard lock(dbg_lock); - std::cerr << rout << ";" << msg<< ";" << c << std::endl; - } - } - PYOPENCL_INLINE const char* - routine() const - { - return m_routine; - } - - PYOPENCL_INLINE cl_int - code() const - { - return m_code; - } - - PYOPENCL_INLINE bool - is_out_of_memory() const - { - // matches Python implementation in pyopencl/cffi_cl.py - return (code() == CL_MEM_OBJECT_ALLOCATION_FAILURE || - code() == CL_OUT_OF_RESOURCES || - code() == CL_OUT_OF_HOST_MEMORY); - } -}; - -// }}} - -// {{{ tracing and error reporting - -template -struct __CLArgGetter { - template - static PYOPENCL_INLINE auto - get(T&& clarg) -> decltype(clarg.convert()) - { - return clarg.convert(); - } -}; - -template -struct __CLFinish { - static PYOPENCL_INLINE void - call(T, bool) - { - } -}; - -template -struct __CLFinish().finish(true)))> { - static PYOPENCL_INLINE void - call(T v, bool converted) - { - v.finish(converted); - } -}; - -template -struct __CLPost { - static PYOPENCL_INLINE void - call(T) - { - } -}; - -template -struct __CLPost().post()))> { - static PYOPENCL_INLINE void - call(T v) - { - v.post(); - } -}; - -template -struct is_out_arg : std::false_type {}; - -template -struct is_out_arg::is_out> > : std::true_type {}; - -template -struct __CLPrintOut { - static PYOPENCL_INLINE void - call(T, std::ostream&) - { - } -}; - -template -struct __CLPrintOut::value> > { - static inline void - call(T v, std::ostream &stm) - { - stm << ", "; - v.print(stm, true); - } -}; - -template -struct __CLPrint { - static inline void - call(T v, std::ostream &stm, bool &&first) - { - if (!first) { - stm << ", "; - } else { - first = false; - } - if (is_out_arg::value) { - stm << "{out}"; - } - v.print(stm); - } -}; - -template class Caller, size_t n, typename T> -struct __CLCall { - template - static PYOPENCL_INLINE void - call(T &&t, Ts&&... ts) - { - __CLCall::call(std::forward(t), - std::forward(ts)...); - Caller(t))>::call(std::get(t), - std::forward(ts)...); - } -}; - -template class Caller, typename T> -struct __CLCall { - template - static PYOPENCL_INLINE void - call(T &&t, Ts&&... ts) - { - Caller(t))>::call(std::get<0>(t), - std::forward(ts)...); - } -}; - -template -class CLArgPack : public ArgPack { - template void - _print_trace(T &res, const char *name) - { - typename CLArgPack::tuple_base *that = this; - std::cerr << name << "("; - __CLCall<__CLPrint, sizeof...(Types) - 1, - decltype(*that)>::call(*that, std::cerr, true); - std::cerr << ") = (ret: " << res; - __CLCall<__CLPrintOut, sizeof...(Types) - 1, - decltype(*that)>::call(*that, std::cerr); - std::cerr << ")" << std::endl; - } -public: - using ArgPack::ArgPack; - template - PYOPENCL_INLINE auto - clcall(Func func, const char *name) - -> decltype(this->template call<__CLArgGetter>(func)) - { - auto res = this->template call<__CLArgGetter>(func); - if (DEBUG_ON) { - std::lock_guard lock(dbg_lock); - _print_trace(res, name); - } - return res; - } - PYOPENCL_INLINE void - finish() - { - typename CLArgPack::tuple_base *that = this; - __CLCall<__CLFinish, sizeof...(Types) - 1, - decltype(*that)>::call(*that, false); - __CLCall<__CLPost, sizeof...(Types) - 1, - decltype(*that)>::call(*that); - __CLCall<__CLFinish, sizeof...(Types) - 1, - decltype(*that)>::call(*that, true); - } -}; - -template -static PYOPENCL_INLINE CLArgPack...> -make_clargpack(Types&&... args) -{ - return CLArgPack...>(std::forward(args)...); -} - -template -static PYOPENCL_INLINE void -call_guarded(cl_int (CL_API_CALL *func)(ArgTypes...), const char *name, ArgTypes2&&... args) -{ - auto argpack = make_clargpack(std::forward(args)...); - cl_int status_code = argpack.clcall(func, name); - if (status_code != CL_SUCCESS) { - throw clerror(name, status_code); - } - argpack.finish(); -} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE T -call_guarded(T (CL_API_CALL *func)(ArgTypes...), const char *name, ArgTypes2&&... args) -{ - cl_int status_code = CL_SUCCESS; - auto status_arg = buf_arg(status_code); - auto argpack = make_clargpack(std::forward(args)..., status_arg); - T res = argpack.clcall(func, name); - if (status_code != CL_SUCCESS) { - throw clerror(name, status_code); - } - argpack.finish(); - return res; -} -#define pyopencl_call_guarded(func, ...) \ - call_guarded(func, #func, __VA_ARGS__) - -static PYOPENCL_INLINE void -cleanup_print_error(cl_int status_code, const char *name) noexcept -{ - std::cerr << ("PyOpenCL WARNING: a clean-up operation failed " - "(dead context maybe?)") << std::endl - << name << " failed with code " << status_code << std::endl; -} - -template -static PYOPENCL_INLINE void -call_guarded_cleanup(cl_int (CL_API_CALL *func)(ArgTypes...), const char *name, - ArgTypes2&&... args) -{ - auto argpack = make_clargpack(std::forward(args)...); - cl_int status_code = argpack.clcall(func, name); - if (status_code != CL_SUCCESS) { - cleanup_print_error(status_code, name); - } else { - argpack.finish(); - } -} -#define pyopencl_call_guarded_cleanup(func, ...) \ - call_guarded_cleanup(func, #func, __VA_ARGS__) - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE error* -c_handle_error(Func func) noexcept -{ - try { - func(); - return nullptr; - } catch (const clerror &e) { - auto err = (error*)malloc(sizeof(error)); - err->routine = strdup(e.routine()); - err->msg = strdup(e.what()); - err->code = e.code(); - err->other = 0; - return err; - } catch (const std::exception &e) { - /* non-pyopencl exceptions need to be converted as well */ - auto err = (error*)malloc(sizeof(error)); - err->other = 1; - err->msg = strdup(e.what()); - return err; - } -} - -template -static PYOPENCL_INLINE auto -retry_mem_error(Func func) -> decltype(func()) -{ - try { - return func(); - } catch (clerror &e) { - if (PYOPENCL_LIKELY(!e.is_out_of_memory()) || !py::gc()) { - throw; - } - } - return func(); -} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE error* -c_handle_retry_mem_error(Func &&func) noexcept -{ - return c_handle_error([&] {retry_mem_error(std::forward(func));}); -} - -// }}} - -#endif diff --git a/src/c_wrapper/event.cpp b/src/c_wrapper/event.cpp deleted file mode 100644 index d75c3a32..00000000 --- a/src/c_wrapper/event.cpp +++ /dev/null @@ -1,294 +0,0 @@ -#include "event.h" -#include "command_queue.h" -#include "context.h" -#include "pyhelper.h" - -#include - -template class clobj; -template void print_arg(std::ostream&, const cl_event&, bool); -template void print_clobj(std::ostream&, const event*); -template void print_buf(std::ostream&, const cl_event*, - size_t, ArgType, bool, bool); - -class event_private { - mutable volatile std::atomic_bool m_finished; - virtual void finish() noexcept = 0; -public: - event_private() - : m_finished(false) - {} - virtual - ~event_private() - {} - void - call_finish() noexcept - { - if (m_finished.exchange(true)) - return; - finish(); - } - bool - is_finished() noexcept - { - return m_finished; - } -}; - -event::event(cl_event event, bool retain, event_private *p) - : clobj(event), m_p(p) -{ - if (retain) { - try { - pyopencl_call_guarded(clRetainEvent, PYOPENCL_CL_CASTABLE_THIS); - } catch (...) { - m_p->call_finish(); - delete m_p; - throw; - } - } -} - -#if PYOPENCL_CL_VERSION >= 0x1010 -static PYOPENCL_INLINE bool -release_private_use_cb(event *evt) -{ - try { - cl_int status = 0; - pyopencl_call_guarded(clGetEventInfo, evt, - CL_EVENT_COMMAND_EXECUTION_STATUS, - size_arg(status), nullptr); - // Event Callback may not be run immediately when the event - // is already completed. - if (status <= CL_COMPLETE) - return false; - cl_context ctx; - pyopencl_call_guarded(clGetEventInfo, evt, CL_EVENT_CONTEXT, - size_arg(ctx), nullptr); - int major; - int minor; - context::get_version(ctx, &major, &minor); - return (major > 1) || (major >= 1 && minor >= 1); - } catch (const clerror &e) { - cleanup_print_error(e.code(), e.what()); - return false; - } -} -#endif - -void -event::release_private() noexcept -{ - if (!m_p) - return; - if (m_p->is_finished()) { - delete m_p; - return; - } -#if PYOPENCL_CL_VERSION >= 0x1010 && defined(PYOPENCL_HAVE_EVENT_SET_CALLBACK) - if (release_private_use_cb(this)) { - try { - event_private *p = m_p; - set_callback(CL_COMPLETE, [p] (cl_int) { - p->call_finish(); - delete p; - }); - return; - } catch (const clerror &e) { - cleanup_print_error(e.code(), e.what()); - } - } -#endif - wait(); - delete m_p; -} - -event::~event() -{ - release_private(); - pyopencl_call_guarded_cleanup(clReleaseEvent, PYOPENCL_CL_CASTABLE_THIS); -} - -generic_info -event::get_info(cl_uint param_name) const -{ - switch ((cl_event_info)param_name) { - case CL_EVENT_COMMAND_QUEUE: - return pyopencl_get_opaque_info(command_queue, Event, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_EVENT_COMMAND_TYPE: - return pyopencl_get_int_info(cl_command_type, Event, - PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_EVENT_COMMAND_EXECUTION_STATUS: - return pyopencl_get_int_info(cl_int, Event, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_EVENT_REFERENCE_COUNT: - return pyopencl_get_int_info(cl_uint, Event, PYOPENCL_CL_CASTABLE_THIS, param_name); -#if PYOPENCL_CL_VERSION >= 0x1010 - case CL_EVENT_CONTEXT: - return pyopencl_get_opaque_info(context, Event, PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif - - default: - throw clerror("Event.get_info", CL_INVALID_VALUE); - } -} - -generic_info -event::get_profiling_info(cl_profiling_info param) const -{ - switch (param) { - case CL_PROFILING_COMMAND_QUEUED: - case CL_PROFILING_COMMAND_SUBMIT: - case CL_PROFILING_COMMAND_START: - case CL_PROFILING_COMMAND_END: - return pyopencl_get_int_info(cl_ulong, EventProfiling, PYOPENCL_CL_CASTABLE_THIS, param); - default: - throw clerror("Event.get_profiling_info", CL_INVALID_VALUE); - } -} - -void -event::wait() const -{ - pyopencl_call_guarded(clWaitForEvents, len_arg(data())); - if (m_p) { - m_p->call_finish(); - } -} - -class nanny_event_private : public event_private { - void *m_ward; - void finish() noexcept - { - void *ward = m_ward; - m_ward = nullptr; - py::deref(ward); - } -public: - nanny_event_private(void *ward) - : m_ward(nullptr) - { - m_ward = py::ref(ward); - } - PYOPENCL_USE_RESULT PYOPENCL_INLINE void* - get_ward() const noexcept - { - return m_ward; - } -}; - -nanny_event::nanny_event(cl_event evt, bool retain, void *ward) - : event(evt, retain, ward ? new nanny_event_private(ward) : nullptr) -{ -} - -PYOPENCL_USE_RESULT void* -nanny_event::get_ward() const noexcept -{ - return (get_p() ? static_cast(get_p())->get_ward() : - nullptr); -} - -#if PYOPENCL_CL_VERSION >= 0x1010 -class user_event : public event { -public: - using event::event; - PYOPENCL_INLINE void - set_status(cl_int status) - { - pyopencl_call_guarded(clSetUserEventStatus, PYOPENCL_CL_CASTABLE_THIS, status); - } -}; -#endif - -// c wrapper - -// Event -error* -event__get_profiling_info(clobj_t _evt, cl_profiling_info param, - generic_info *out) -{ - auto evt = static_cast(_evt); - return c_handle_error([&] { - *out = evt->get_profiling_info(param); - }); -} - -error* -event__wait(clobj_t evt) -{ - return c_handle_error([&] { - static_cast(evt)->wait(); - }); -} - - -error* -event__set_callback(clobj_t _evt, cl_int type, void *pyobj) -{ -#if PYOPENCL_CL_VERSION >= 0x1010 && defined(PYOPENCL_HAVE_EVENT_SET_CALLBACK) - auto evt = static_cast(_evt); - return c_handle_error([&] { - pyobj = py::ref(pyobj); - try { - evt->set_callback(type, [=] (cl_int status) { - py::call(pyobj, status); - py::deref(pyobj); - }); - } catch (...) { - py::deref(pyobj); - } - }); -#else - PYOPENCL_UNSUPPORTED(clSetEventCallback, "CL 1.0 and below and Windows") -#endif -} - -// Nanny Event -void* -nanny_event__get_ward(clobj_t evt) -{ - return static_cast(evt)->get_ward(); -} - -error* -wait_for_events(const clobj_t *_wait_for, uint32_t num_wait_for) -{ - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - return c_handle_error([&] { - pyopencl_call_guarded(clWaitForEvents, wait_for); - }); -} - -error* -enqueue_wait_for_events(clobj_t _queue, const clobj_t *_wait_for, - uint32_t num_wait_for) -{ - auto queue = static_cast(_queue); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - return c_handle_error([&] { - pyopencl_call_guarded(clEnqueueWaitForEvents, queue, wait_for); - }); -} - -#if PYOPENCL_CL_VERSION >= 0x1010 - -error* -create_user_event(clobj_t *_evt, clobj_t _ctx) -{ - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - auto evt = pyopencl_call_guarded(clCreateUserEvent, ctx); - *_evt = pyopencl_convert_obj(user_event, clReleaseEvent, evt); - }); -} - -error* -user_event__set_status(clobj_t _evt, cl_int status) -{ - auto evt = static_cast(_evt); - return c_handle_error([&] { - evt->set_status(status); - }); -} - -#endif diff --git a/src/c_wrapper/event.h b/src/c_wrapper/event.h deleted file mode 100644 index c6d0dd4b..00000000 --- a/src/c_wrapper/event.h +++ /dev/null @@ -1,87 +0,0 @@ -#include "clhelper.h" -#include - -#ifndef __PYOPENCL_EVENT_H -#define __PYOPENCL_EVENT_H - -// {{{ event - -extern template class clobj; -extern template void print_arg(std::ostream&, const cl_event&, bool); -extern template void print_buf(std::ostream&, const cl_event*, - size_t, ArgType, bool, bool); - -class event_private; - -class event : public clobj { - event_private *m_p; - // return whether the event need to be released. - void release_private() noexcept; -protected: - PYOPENCL_INLINE event_private* - get_p() const - { - return m_p; - } -public: - PYOPENCL_DEF_CL_CLASS(EVENT); - event(cl_event event, bool retain, event_private *p=nullptr); - ~event(); - generic_info get_info(cl_uint param) const; - PYOPENCL_USE_RESULT generic_info - get_profiling_info(cl_profiling_info param) const; - void wait() const; -#if PYOPENCL_CL_VERSION >= 0x1010 && defined(PYOPENCL_HAVE_EVENT_SET_CALLBACK) - template - PYOPENCL_INLINE void - set_callback(cl_int type, Func &&_func) - { - auto func = new rm_ref_t(std::forward(_func)); - try { - pyopencl_call_guarded( - clSetEventCallback, PYOPENCL_CL_CASTABLE_THIS, type, - static_cast( - [] (cl_event, cl_int status, void *data) { - rm_ref_t *func = static_cast*>(data); - - // We won't necessarily be able to acquire the GIL inside this - // handler without deadlocking. Create a thread that *can* - // wait. - - std::thread t([func, status] () { - (*func)(status); - delete func; - }); - t.detach(); - - }), (void*)func); - } catch (...) { - delete func; - throw; - } - } -#endif -}; -static PYOPENCL_INLINE auto -event_out(clobj_t *ret) -> decltype(pyopencl_outarg(event, ret, clReleaseEvent)) -{ - return pyopencl_outarg(event, ret, clReleaseEvent); -} - -extern template void print_clobj(std::ostream&, const event*); - -class nanny_event : public event { -public: - nanny_event(cl_event evt, bool retain, void *ward=nullptr); - PYOPENCL_USE_RESULT void *get_ward() const noexcept; -}; -static PYOPENCL_INLINE auto -nanny_event_out(clobj_t *ret, void *ward) - -> decltype(pyopencl_outarg(nanny_event, ret, clReleaseEvent, ward)) -{ - return pyopencl_outarg(nanny_event, ret, clReleaseEvent, ward); -} - -// }}} - -#endif diff --git a/src/c_wrapper/function.h b/src/c_wrapper/function.h deleted file mode 100644 index 5d1a604c..00000000 --- a/src/c_wrapper/function.h +++ /dev/null @@ -1,121 +0,0 @@ -#include -#include - -#ifndef __PYOPENCL_FUNCTION_H -#define __PYOPENCL_FUNCTION_H - -#if defined __GNUC__ && __GNUC__ > 3 -#define PYOPENCL_INLINE inline __attribute__((__always_inline__)) -#else -#define PYOPENCL_INLINE inline -#endif - -template -using rm_ref_t = typename std::remove_reference::type; -template -using rm_const_t = typename std::remove_const::type; -template -using enable_if_t = typename std::enable_if::type; - -template -struct seq { -}; - -template -struct gens : gens { -}; - -template -struct gens<0, S...> { - typedef seq type; -}; - -template -static PYOPENCL_INLINE auto -_call_func(Function func, seq, std::tuple &args) - -> decltype(func(std::forward(std::get(args))...)) -{ - return func(static_cast(std::get(args))...); -} - -template -static PYOPENCL_INLINE auto -call_tuple(Function &&func, T &&args) - -> decltype(_call_func(std::forward(func), - typename gens::value>::type(), - args)) -{ - return _call_func(std::forward(func), - typename gens::value>::type(), args); -} - -template class Convert, typename... Types> -using _ArgPackBase = std::tuple::type>...>; - -template class Convert, typename... Types> -class ArgPack : public _ArgPackBase { -public: - typedef _ArgPackBase tuple_base; -private: - template - static PYOPENCL_INLINE std::tuple - ensure_tuple(T &&v) - { - return std::tuple(std::forward(v)); - } - template - static PYOPENCL_INLINE std::tuple - ensure_tuple(std::tuple &&t) - { - return t; - } - - template - using ArgConvert = Convert >; - template class Getter, int... S> - PYOPENCL_INLINE auto - __get(seq) -#ifndef _MSC_VER - -> decltype(std::tuple_cat( - ensure_tuple(Getter >::get( - std::get(*(tuple_base*)this)))...)) -#endif - { - return std::tuple_cat( - ensure_tuple(Getter >::get( - std::get(*(tuple_base*)this)))...); - } -public: - template - ArgPack(Types2&&... arg_orig) - : tuple_base(ArgConvert >(arg_orig)...) - { - } - ArgPack(ArgPack &&other) - : tuple_base(static_cast(other)) - { - } - // GCC Bug: https://gcc.gnu.org/bugzilla/show_bug.cgi?id=57543 - template class Getter> - PYOPENCL_INLINE auto - get() -> decltype(this->__get( - typename gens::type())) - { - return __get(typename gens::type()); - } - template class Getter, typename Func> - PYOPENCL_INLINE auto - call(Func func) -> decltype(call_tuple(func, this->get())) - { - return call_tuple(func, this->get()); - } -}; - -template class Convert, typename... Types> -static PYOPENCL_INLINE ArgPack...> -make_argpack(Types&&... args) -{ - return ArgPack...>(std::forward(args)...); -} - -#endif diff --git a/src/c_wrapper/gl_obj.cpp b/src/c_wrapper/gl_obj.cpp deleted file mode 100644 index bd7edf31..00000000 --- a/src/c_wrapper/gl_obj.cpp +++ /dev/null @@ -1,155 +0,0 @@ -#include "gl_obj.h" -#include "context.h" -#include "command_queue.h" -#include "event.h" -#include "clhelper.h" - -#ifdef HAVE_GL - -template void print_clobj(std::ostream&, const gl_buffer*); -template void print_clobj(std::ostream&, - const gl_renderbuffer*); - -generic_info -gl_texture::get_gl_texture_info(cl_gl_texture_info param_name) const -{ - switch (param_name) { - case CL_GL_TEXTURE_TARGET: - return pyopencl_get_int_info(GLenum, GLTexture, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_GL_MIPMAP_LEVEL: - return pyopencl_get_int_info(GLint, GLTexture, PYOPENCL_CL_CASTABLE_THIS, param_name); - default: - throw clerror("MemoryObject.get_gl_texture_info", CL_INVALID_VALUE); - } -} - -typedef cl_int (CL_API_CALL *clEnqueueGLObjectFunc)(cl_command_queue, cl_uint, - const cl_mem*, cl_uint, - const cl_event*, cl_event*); - -static PYOPENCL_INLINE void -enqueue_gl_objects(clEnqueueGLObjectFunc func, const char *name, - clobj_t *evt, command_queue *cq, const clobj_t *mem_objects, - uint32_t num_mem_objects, const clobj_t *wait_for, - uint32_t num_wait_for) -{ - const auto _wait_for = buf_from_class(wait_for, num_wait_for); - const auto _mem_objs = buf_from_class( - mem_objects, num_mem_objects); - call_guarded(func, name, cq, _mem_objs, _wait_for, event_out(evt)); -} -#define enqueue_gl_objects(what, ...) \ - enqueue_gl_objects(clEnqueue##what##GLObjects, \ - "clEnqueue" #what "GLObjects", __VA_ARGS__) - -// c wrapper - -error* -create_from_gl_texture(clobj_t *ptr, clobj_t _ctx, cl_mem_flags flags, - GLenum texture_target, GLint miplevel, - GLuint texture) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - cl_mem mem = pyopencl_call_guarded(clCreateFromGLTexture, - ctx, flags, texture_target, miplevel, texture); - *ptr = pyopencl_convert_obj(gl_texture, clReleaseMemObject, mem); - }); -#else - PYOPENCL_UNSUPPORTED(clCreateFromGLTexture, "CL 1.1") -#endif -} - -error* -create_from_gl_buffer(clobj_t *ptr, clobj_t _ctx, - cl_mem_flags flags, GLuint bufobj) -{ - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - cl_mem mem = pyopencl_call_guarded(clCreateFromGLBuffer, - ctx, flags, bufobj); - *ptr = pyopencl_convert_obj(gl_buffer, clReleaseMemObject, mem); - }); -} - -error* -create_from_gl_renderbuffer(clobj_t *ptr, clobj_t _ctx, - cl_mem_flags flags, GLuint bufobj) -{ - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - cl_mem mem = pyopencl_call_guarded(clCreateFromGLRenderbuffer, - ctx, flags, bufobj); - *ptr = pyopencl_convert_obj(gl_renderbuffer, - clReleaseMemObject, mem); - }); -} - -error* -enqueue_acquire_gl_objects(clobj_t *evt, clobj_t queue, - const clobj_t *mem_objects, - uint32_t num_mem_objects, - const clobj_t *wait_for, uint32_t num_wait_for) -{ - return c_handle_error([&] { - enqueue_gl_objects( - Acquire, evt, static_cast(queue), - mem_objects, num_mem_objects, wait_for, num_wait_for); - }); -} - -error* -enqueue_release_gl_objects(clobj_t *evt, clobj_t queue, - const clobj_t *mem_objects, - uint32_t num_mem_objects, - const clobj_t *wait_for, uint32_t num_wait_for) -{ - return c_handle_error([&] { - enqueue_gl_objects( - Release, evt, static_cast(queue), - mem_objects, num_mem_objects, wait_for, num_wait_for); - }); -} - -error* -get_gl_object_info(clobj_t mem, cl_gl_object_type *otype, GLuint *gl_name) -{ - auto globj = static_cast(mem); - return c_handle_error([&] { - pyopencl_call_guarded(clGetGLObjectInfo, globj, buf_arg(*otype), - buf_arg(*gl_name)); - }); -} - -#endif - -int -have_gl() -{ -#ifdef HAVE_GL - return 1; -#else - return 0; -#endif -} - -cl_context_properties -get_apple_cgl_share_group() -{ -#if (defined(__APPLE__) && !defined(PYOPENCL_APPLE_USE_CL_H)) - #ifdef HAVE_GL - CGLContextObj kCGLContext = CGLGetCurrentContext(); - CGLShareGroupObj kCGLShareGroup = CGLGetShareGroup(kCGLContext); - - return (cl_context_properties)kCGLShareGroup; - #else - throw clerror("get_apple_cgl_share_group unavailable: " - "GL interop not compiled", - CL_INVALID_VALUE); - #endif -#else - throw clerror("get_apple_cgl_share_group unavailable: non-Apple platform", - CL_INVALID_VALUE); -#endif /* __APPLE__ */ -} diff --git a/src/c_wrapper/gl_obj.h b/src/c_wrapper/gl_obj.h deleted file mode 100644 index 9f47e19b..00000000 --- a/src/c_wrapper/gl_obj.h +++ /dev/null @@ -1,46 +0,0 @@ -#include "image.h" - -#ifndef __PYOPENCL_GL_OBJ_H -#define __PYOPENCL_GL_OBJ_H - -#ifdef HAVE_GL - -// {{{ gl interop - -class gl_buffer : public memory_object { -public: - PYOPENCL_DEF_CL_CLASS(GL_BUFFER); - PYOPENCL_INLINE - gl_buffer(cl_mem mem, bool retain) - : memory_object(mem, retain) - {} -}; - -class gl_renderbuffer : public memory_object { -public: - PYOPENCL_DEF_CL_CLASS(GL_RENDERBUFFER); - PYOPENCL_INLINE - gl_renderbuffer(cl_mem mem, bool retain) - : memory_object(mem, retain) - {} -}; - -extern template void print_clobj(std::ostream&, const gl_buffer*); -extern template void print_clobj(std::ostream&, - const gl_renderbuffer*); - -class gl_texture : public image { - public: - PYOPENCL_INLINE - gl_texture(cl_mem mem, bool retain) - : image(mem, retain) - {} - PYOPENCL_USE_RESULT generic_info - get_gl_texture_info(cl_gl_texture_info param_name) const; -}; - -// }}} - -#endif - -#endif diff --git a/src/c_wrapper/image.cpp b/src/c_wrapper/image.cpp deleted file mode 100644 index 6f571f32..00000000 --- a/src/c_wrapper/image.cpp +++ /dev/null @@ -1,237 +0,0 @@ -#include "image.h" -#include "context.h" -#include "command_queue.h" -#include "event.h" -#include "buffer.h" - -template void print_clobj(std::ostream&, const image*); - -PYOPENCL_USE_RESULT static PYOPENCL_INLINE image* -new_image(cl_mem mem, const cl_image_format *fmt) -{ - return pyopencl_convert_obj(image, clReleaseMemObject, mem, fmt); -} - -generic_info -image::get_image_info(cl_image_info param) const -{ - switch (param) { - case CL_IMAGE_FORMAT: - return pyopencl_get_int_info(cl_image_format, Image, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_IMAGE_ELEMENT_SIZE: - case CL_IMAGE_ROW_PITCH: - case CL_IMAGE_SLICE_PITCH: - case CL_IMAGE_WIDTH: - case CL_IMAGE_HEIGHT: - case CL_IMAGE_DEPTH: -#if PYOPENCL_CL_VERSION >= 0x1020 - case CL_IMAGE_ARRAY_SIZE: -#endif - return pyopencl_get_int_info(size_t, Image, PYOPENCL_CL_CASTABLE_THIS, param); - -#if PYOPENCL_CL_VERSION >= 0x1020 - // TODO: - // case CL_IMAGE_BUFFER: - // { - // cl_mem param_value; - // PYOPENCL_CALL_GUARDED(clGetImageInfo, (this, param, sizeof(param_value), ¶m_value, 0)); - // if (param_value == 0) - // { - // // no associated memory object? no problem. - // return py::object(); - // } - // return create_mem_object_wrapper(param_value); - // } - case CL_IMAGE_NUM_MIP_LEVELS: - case CL_IMAGE_NUM_SAMPLES: - return pyopencl_get_int_info(cl_uint, Image, PYOPENCL_CL_CASTABLE_THIS, param); -#endif - default: - throw clerror("Image.get_image_info", CL_INVALID_VALUE); - } -} - -// c wrapper - -// Image -error* -create_image_2d(clobj_t *img, clobj_t _ctx, cl_mem_flags flags, - cl_image_format *fmt, size_t width, size_t height, - size_t pitch, void *buf) -{ - auto ctx = static_cast(_ctx); - return c_handle_retry_mem_error([&] { - auto mem = pyopencl_call_guarded(clCreateImage2D, ctx, flags, fmt, - width, height, pitch, buf); - *img = new_image(mem, fmt); - }); -} - -error* -create_image_3d(clobj_t *img, clobj_t _ctx, cl_mem_flags flags, - cl_image_format *fmt, size_t width, size_t height, - size_t depth, size_t pitch_x, size_t pitch_y, void *buf) -{ - auto ctx = static_cast(_ctx); - return c_handle_retry_mem_error([&] { - auto mem = pyopencl_call_guarded(clCreateImage3D, ctx, flags, fmt, - width, height, depth, pitch_x, - pitch_y, buf); - *img = new_image(mem, fmt); - }); -} - - -error* -create_image_from_desc(clobj_t *img, clobj_t _ctx, cl_mem_flags flags, - cl_image_format *fmt, cl_image_desc *desc, void *buf) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - auto mem = pyopencl_call_guarded(clCreateImage, ctx, flags, fmt, - desc, buf); - *img = new_image(mem, fmt); - }); -#else - PYOPENCL_UNSUPPORTED(clCreateImage, "CL 1.1 and below") -#endif -} - - -error* -image__get_image_info(clobj_t _img, cl_image_info param, generic_info *out) -{ - auto img = static_cast(_img); - return c_handle_error([&] { - *out = img->get_image_info(param); - }); -} - -type_t -image__get_fill_type(clobj_t img) -{ - return static_cast(img)->get_fill_type(); -} - -error* -enqueue_read_image(clobj_t *evt, clobj_t _queue, clobj_t _mem, - const size_t *_orig, size_t orig_l, - const size_t *_reg, size_t reg_l, void *buf, - size_t row_pitch, size_t slice_pitch, - const clobj_t *_wait_for, uint32_t num_wait_for, - int block, void *pyobj) -{ - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - auto img = static_cast(_mem); - ConstBuffer orig(_orig, orig_l); - ConstBuffer reg(_reg, reg_l, 1); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueReadImage, queue, img, bool(block), - orig, reg, row_pitch, slice_pitch, buf, - wait_for, nanny_event_out(evt, pyobj)); - }); -} - -error* -enqueue_copy_image(clobj_t *evt, clobj_t _queue, clobj_t _src, clobj_t _dst, - const size_t *_src_orig, size_t src_orig_l, - const size_t *_dst_orig, size_t dst_orig_l, - const size_t *_reg, size_t reg_l, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - auto src = static_cast(_src); - auto dst = static_cast(_dst); - ConstBuffer src_orig(_src_orig, src_orig_l); - ConstBuffer dst_orig(_dst_orig, dst_orig_l); - ConstBuffer reg(_reg, reg_l, 1); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueCopyImage, queue, src, dst, src_orig, - dst_orig, reg, wait_for, event_out(evt)); - }); -} - -error* -enqueue_write_image(clobj_t *evt, clobj_t _queue, clobj_t _mem, - const size_t *_orig, size_t orig_l, - const size_t *_reg, size_t reg_l, - const void *buf, size_t row_pitch, size_t slice_pitch, - const clobj_t *_wait_for, uint32_t num_wait_for, - int block, void *pyobj) -{ - auto queue = static_cast(_queue); - auto img = static_cast(_mem); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - ConstBuffer orig(_orig, orig_l); - ConstBuffer reg(_reg, reg_l, 1); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueWriteImage, queue, img, bool(block), - orig, reg, row_pitch, slice_pitch, buf, - wait_for, nanny_event_out(evt, pyobj)); - }); -} - -error* -enqueue_fill_image(clobj_t *evt, clobj_t _queue, clobj_t mem, - const void *color, const size_t *_orig, size_t orig_l, - const size_t *_reg, size_t reg_l, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - // TODO debug color - auto queue = static_cast(_queue); - auto img = static_cast(mem); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - ConstBuffer orig(_orig, orig_l); - ConstBuffer reg(_reg, reg_l, 1); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueFillImage, queue, img, color, orig, - reg, wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED(clEnqueueFillImage, "CL 1.1 and below") -#endif -} - -// {{{ image transfers - -error* -enqueue_copy_image_to_buffer(clobj_t *evt, clobj_t _queue, clobj_t _src, - clobj_t _dst, const size_t *_orig, size_t orig_l, - const size_t *_reg, size_t reg_l, size_t offset, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ - auto queue = static_cast(_queue); - auto src = static_cast(_src); - auto dst = static_cast(_dst); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - ConstBuffer orig(_orig, orig_l); - ConstBuffer reg(_reg, reg_l, 1); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueCopyImageToBuffer, queue, src, dst, - orig, reg, offset, wait_for, event_out(evt)); - }); -} - -error* -enqueue_copy_buffer_to_image(clobj_t *evt, clobj_t _queue, clobj_t _src, - clobj_t _dst, size_t offset, const size_t *_orig, - size_t orig_l, const size_t *_reg, size_t reg_l, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ - auto queue = static_cast(_queue); - auto src = static_cast(_src); - auto dst = static_cast(_dst); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - ConstBuffer orig(_orig, orig_l); - ConstBuffer reg(_reg, reg_l, 1); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueCopyBufferToImage, queue, src, dst, - offset, orig, reg, wait_for, event_out(evt)); - }); -} - -// }}} diff --git a/src/c_wrapper/image.h b/src/c_wrapper/image.h deleted file mode 100644 index 7d29909c..00000000 --- a/src/c_wrapper/image.h +++ /dev/null @@ -1,50 +0,0 @@ -#include "memory_object.h" -#include "clhelper.h" - -#ifndef __PYOPENCL_IMAGE_H -#define __PYOPENCL_IMAGE_H - -// {{{ image - -class image : public memory_object { -private: - cl_image_format m_format; -public: - PYOPENCL_DEF_CL_CLASS(IMAGE); - PYOPENCL_INLINE - image(cl_mem mem, bool retain, const cl_image_format *fmt=0) - : memory_object(mem, retain), m_format(fmt ? *fmt : cl_image_format()) - {} - PYOPENCL_INLINE const cl_image_format& - format() - { - if (!m_format.image_channel_data_type) { - pyopencl_call_guarded(clGetImageInfo, PYOPENCL_CL_CASTABLE_THIS, CL_IMAGE_FORMAT, - size_arg(m_format), nullptr); - } - return m_format; - } - PYOPENCL_USE_RESULT generic_info get_image_info(cl_image_info param) const; - PYOPENCL_INLINE type_t - get_fill_type() - { - switch (format().image_channel_data_type) { - case CL_SIGNED_INT8: - case CL_SIGNED_INT16: - case CL_SIGNED_INT32: - return TYPE_INT; - case CL_UNSIGNED_INT8: - case CL_UNSIGNED_INT16: - case CL_UNSIGNED_INT32: - return TYPE_UINT; - default: - return TYPE_FLOAT; - } - } -}; - -extern template void print_clobj(std::ostream&, const image*); - -// }}} - -#endif diff --git a/src/c_wrapper/kernel.cpp b/src/c_wrapper/kernel.cpp deleted file mode 100644 index 817e1061..00000000 --- a/src/c_wrapper/kernel.cpp +++ /dev/null @@ -1,213 +0,0 @@ -#include "kernel.h" -#include "context.h" -#include "device.h" -#include "program.h" -#include "memory_object.h" -#include "sampler.h" -#include "command_queue.h" -#include "event.h" -#include "clhelper.h" - -template class clobj; -template void print_arg(std::ostream&, const cl_kernel&, bool); -template void print_clobj(std::ostream&, const kernel*); -template void print_buf(std::ostream&, const cl_kernel*, - size_t, ArgType, bool, bool); - -kernel::~kernel() -{ - pyopencl_call_guarded_cleanup(clReleaseKernel, PYOPENCL_CL_CASTABLE_THIS); -} - -generic_info -kernel::get_info(cl_uint param) const -{ - switch ((cl_kernel_info)param) { - case CL_KERNEL_FUNCTION_NAME: - return pyopencl_get_str_info(Kernel, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_KERNEL_NUM_ARGS: - case CL_KERNEL_REFERENCE_COUNT: - return pyopencl_get_int_info(cl_uint, Kernel, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_KERNEL_CONTEXT: - return pyopencl_get_opaque_info(context, Kernel, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_KERNEL_PROGRAM: - return pyopencl_get_opaque_info(program, Kernel, PYOPENCL_CL_CASTABLE_THIS, param); -#if PYOPENCL_CL_VERSION >= 0x1020 - case CL_KERNEL_ATTRIBUTES: - return pyopencl_get_str_info(Kernel, PYOPENCL_CL_CASTABLE_THIS, param); -#endif - default: - throw clerror("Kernel.get_info", CL_INVALID_VALUE); - } -} - -generic_info -kernel::get_work_group_info(cl_kernel_work_group_info param, - const device *dev) const -{ - switch (param) { -#if PYOPENCL_CL_VERSION >= 0x1010 - case CL_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE: -#endif - case CL_KERNEL_WORK_GROUP_SIZE: - return pyopencl_get_int_info(size_t, KernelWorkGroup, PYOPENCL_CL_CASTABLE_THIS, dev, param); - case CL_KERNEL_COMPILE_WORK_GROUP_SIZE: - return pyopencl_get_array_info(size_t, KernelWorkGroup, - PYOPENCL_CL_CASTABLE_THIS, dev, param); - case CL_KERNEL_LOCAL_MEM_SIZE: -#if PYOPENCL_CL_VERSION >= 0x1010 - case CL_KERNEL_PRIVATE_MEM_SIZE: -#endif - return pyopencl_get_int_info(cl_ulong, KernelWorkGroup, - PYOPENCL_CL_CASTABLE_THIS, dev, param); - default: - throw clerror("Kernel.get_work_group_info", CL_INVALID_VALUE); - } -} - -#if PYOPENCL_CL_VERSION >= 0x1020 -PYOPENCL_USE_RESULT generic_info -kernel::get_arg_info(cl_uint idx, cl_kernel_arg_info param) const -{ - switch (param) { - case CL_KERNEL_ARG_ADDRESS_QUALIFIER: - return pyopencl_get_int_info(cl_kernel_arg_address_qualifier, - KernelArg, PYOPENCL_CL_CASTABLE_THIS, idx, param); - case CL_KERNEL_ARG_ACCESS_QUALIFIER: - return pyopencl_get_int_info(cl_kernel_arg_access_qualifier, - KernelArg, PYOPENCL_CL_CASTABLE_THIS, idx, param); - case CL_KERNEL_ARG_TYPE_QUALIFIER: - return pyopencl_get_int_info(cl_kernel_arg_type_qualifier, - KernelArg, PYOPENCL_CL_CASTABLE_THIS, idx, param); - case CL_KERNEL_ARG_TYPE_NAME: - case CL_KERNEL_ARG_NAME: - return pyopencl_get_str_info(KernelArg, PYOPENCL_CL_CASTABLE_THIS, idx, param); - default: - throw clerror("Kernel.get_arg_info", CL_INVALID_VALUE); - } -} -#endif - -// c wrapper - -// Kernel -error* -create_kernel(clobj_t *knl, clobj_t _prog, const char *name) -{ - auto prog = static_cast(_prog); - return c_handle_error([&] { - *knl = new kernel(pyopencl_call_guarded(clCreateKernel, prog, - name), false); - }); -} - -error* -kernel__set_arg_null(clobj_t _knl, cl_uint arg_index) -{ - auto knl = static_cast(_knl); - return c_handle_error([&] { - const cl_mem m = 0; - pyopencl_call_guarded(clSetKernelArg, knl, arg_index, size_arg(m)); - }); -} - -error* -kernel__set_arg_mem(clobj_t _knl, cl_uint arg_index, clobj_t _mem) -{ - auto knl = static_cast(_knl); - auto mem = static_cast(_mem); - return c_handle_error([&] { - pyopencl_call_guarded(clSetKernelArg, knl, arg_index, - size_arg(mem->data())); - }); -} - -error* -kernel__set_arg_sampler(clobj_t _knl, cl_uint arg_index, clobj_t _samp) -{ - auto knl = static_cast(_knl); - auto samp = static_cast(_samp); - return c_handle_error([&] { - pyopencl_call_guarded(clSetKernelArg, knl, arg_index, - size_arg(samp->data())); - }); -} - -error* -kernel__set_arg_buf(clobj_t _knl, cl_uint arg_index, - const void *buffer, size_t size) -{ - auto knl = static_cast(_knl); - return c_handle_error([&] { - pyopencl_call_guarded(clSetKernelArg, knl, arg_index, - size_arg(buffer, size)); - }); -} - -error* -kernel__set_arg_svm_pointer(clobj_t _knl, cl_uint arg_index, void *value) -{ -#if PYOPENCL_CL_VERSION >= 0x2000 - auto knl = static_cast(_knl); - return c_handle_error([&] { - pyopencl_call_guarded(clSetKernelArgSVMPointer, knl, arg_index, value); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clSetKernelArgSVMPointer, "CL 2.0") -#endif -} - -error* -kernel__get_work_group_info(clobj_t _knl, cl_kernel_work_group_info param, - clobj_t _dev, generic_info *out) -{ - auto knl = static_cast(_knl); - auto dev = static_cast(_dev); - return c_handle_error([&] { - *out = knl->get_work_group_info(param, dev); - }); -} - -error* -kernel__get_arg_info(clobj_t _knl, cl_uint idx, cl_kernel_arg_info param, - generic_info *out) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - auto knl = static_cast(_knl); - return c_handle_error([&] { - *out = knl->get_arg_info(idx, param); - }); -#else - PYOPENCL_UNSUPPORTED(clKernelGetArgInfo, "CL 1.1 and below") -#endif -} - -error* -enqueue_nd_range_kernel(clobj_t *evt, clobj_t _queue, clobj_t _knl, - cl_uint work_dim, const size_t *global_work_offset, - const size_t *global_work_size, - const size_t *local_work_size, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ - auto queue = static_cast(_queue); - auto knl = static_cast(_knl); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueNDRangeKernel, queue, knl, work_dim, - global_work_offset, global_work_size, - local_work_size, wait_for, event_out(evt)); - }); -} - -error* -enqueue_task(clobj_t *evt, clobj_t _queue, clobj_t _knl, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ - auto queue = static_cast(_queue); - auto knl = static_cast(_knl); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueTask, queue, knl, wait_for, - event_out(evt)); - }); -} diff --git a/src/c_wrapper/kernel.h b/src/c_wrapper/kernel.h deleted file mode 100644 index 5db1a0cc..00000000 --- a/src/c_wrapper/kernel.h +++ /dev/null @@ -1,44 +0,0 @@ -#include "error.h" - -#ifndef __PYOPENCL_KERNEL_H -#define __PYOPENCL_KERNEL_H - -class device; - -// {{{ kernel - -extern template class clobj; -extern template void print_arg(std::ostream&, - const cl_kernel&, bool); -extern template void print_buf(std::ostream&, const cl_kernel*, - size_t, ArgType, bool, bool); - -class kernel : public clobj { -public: - PYOPENCL_DEF_CL_CLASS(KERNEL); - PYOPENCL_INLINE - kernel(cl_kernel knl, bool retain) - : clobj(knl) - { - if (retain) { - pyopencl_call_guarded(clRetainKernel, PYOPENCL_CL_CASTABLE_THIS); - } - } - ~kernel(); - generic_info get_info(cl_uint param) const; - - PYOPENCL_USE_RESULT generic_info - get_work_group_info(cl_kernel_work_group_info param, - const device *dev) const; - -#if PYOPENCL_CL_VERSION >= 0x1020 - PYOPENCL_USE_RESULT generic_info - get_arg_info(cl_uint idx, cl_kernel_arg_info param) const; -#endif -}; - -extern template void print_clobj(std::ostream&, const kernel*); - -// }}} - -#endif diff --git a/src/c_wrapper/memory_map.cpp b/src/c_wrapper/memory_map.cpp deleted file mode 100644 index 068274df..00000000 --- a/src/c_wrapper/memory_map.cpp +++ /dev/null @@ -1,115 +0,0 @@ -#include "memory_map.h" -#include "image.h" -#include "buffer.h" -#include "event.h" -#include "clhelper.h" - -template class clobj; -template void print_arg(std::ostream&, void *const&, bool); -template void print_buf(std::ostream&, void *const*, - size_t, ArgType, bool, bool); - -memory_map::~memory_map() -{ - if (!m_valid.exchange(false)) - return; - pyopencl_call_guarded_cleanup(clEnqueueUnmapMemObject, m_queue, - m_mem, PYOPENCL_CL_CASTABLE_THIS, 0, nullptr, nullptr); -} - -void -memory_map::release(clobj_t *evt, const command_queue *queue, - const clobj_t *_wait_for, uint32_t num_wait_for) const -{ - if (!m_valid.exchange(false)) { - throw clerror("MemoryMap.release", CL_INVALID_VALUE, - "trying to double-unref mem map"); - } - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - queue = queue ? queue : &m_queue; - pyopencl_call_guarded(clEnqueueUnmapMemObject, queue, - m_mem, PYOPENCL_CL_CASTABLE_THIS, wait_for, event_out(evt)); -} - -generic_info -memory_map::get_info(cl_uint) const -{ - throw clerror("MemoryMap.get_info", CL_INVALID_VALUE); -} - -intptr_t -memory_map::intptr() const -{ - return m_valid ? (intptr_t)data() : 0; -} - -memory_map* -convert_memory_map(clobj_t evt, command_queue *queue, - memory_object *buf, void *res) -{ - try { - return new memory_map(queue, buf, res); - } catch (...) { - delete evt; - pyopencl_call_guarded_cleanup(clEnqueueUnmapMemObject, queue, - buf, res, 0, nullptr, nullptr); - throw; - } -} - -// c wrapper - -// Memory Map -error* -memory_map__release(clobj_t _map, clobj_t _queue, const clobj_t *_wait_for, - uint32_t num_wait_for, clobj_t *evt) -{ - auto map = static_cast(_map); - auto queue = static_cast(_queue); - return c_handle_error([&] { - map->release(evt, queue, _wait_for, num_wait_for); - }); -} - -void* -memory_map__data(clobj_t _map) -{ - return static_cast(_map)->data(); -} - -error* -enqueue_map_image(clobj_t *evt, clobj_t *map, clobj_t _queue, clobj_t _mem, - cl_map_flags flags, const size_t *_orig, size_t orig_l, - const size_t *_reg, size_t reg_l, size_t *row_pitch, - size_t *slice_pitch, const clobj_t *_wait_for, - uint32_t num_wait_for, int block) -{ - auto queue = static_cast(_queue); - auto img = static_cast(_mem); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - ConstBuffer orig(_orig, orig_l); - ConstBuffer reg(_reg, reg_l, 1); - return c_handle_retry_mem_error([&] { - void *res = pyopencl_call_guarded( - clEnqueueMapImage, queue, img, bool(block), flags, orig, - reg, row_pitch, slice_pitch, wait_for, event_out(evt)); - *map = convert_memory_map(*evt, queue, img, res); - }); -} - -error* -enqueue_map_buffer(clobj_t *evt, clobj_t *map, clobj_t _queue, clobj_t _mem, - cl_map_flags flags, size_t offset, size_t size, - const clobj_t *_wait_for, uint32_t num_wait_for, - int block) -{ - auto queue = static_cast(_queue); - auto buf = static_cast(_mem); - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - return c_handle_retry_mem_error([&] { - void *res = pyopencl_call_guarded( - clEnqueueMapBuffer, queue, buf, bool(block), - flags, offset, size, wait_for, event_out(evt)); - *map = convert_memory_map(*evt, queue, buf, res); - }); -} diff --git a/src/c_wrapper/memory_map.h b/src/c_wrapper/memory_map.h deleted file mode 100644 index 65a988a9..00000000 --- a/src/c_wrapper/memory_map.h +++ /dev/null @@ -1,37 +0,0 @@ -#include "error.h" -#include "command_queue.h" -#include "memory_object.h" - -#ifndef __PYOPENCL_MEMORY_MAP_H -#define __PYOPENCL_MEMORY_MAP_H - -class event; - -// {{{ memory_map - -extern template class clobj; -extern template void print_arg(std::ostream&, void *const&, bool); -extern template void print_buf(std::ostream&, void *const*, - size_t, ArgType, bool, bool); - -class memory_map : public clobj { -private: - mutable volatile std::atomic_bool m_valid; - command_queue m_queue; - memory_object m_mem; -public: - constexpr static const char *class_name = "MEMORY_MAP"; - PYOPENCL_INLINE - memory_map(const command_queue *queue, const memory_object *mem, void *ptr) - : clobj(ptr), m_valid(true), m_queue(*queue), m_mem(*mem) - {} - ~memory_map(); - void release(clobj_t *evt, const command_queue *queue, - const clobj_t *wait_for, uint32_t num_wait_for) const; - generic_info get_info(cl_uint) const; - intptr_t intptr() const; -}; - -// }}} - -#endif diff --git a/src/c_wrapper/memory_object.cpp b/src/c_wrapper/memory_object.cpp deleted file mode 100644 index 6f1ba321..00000000 --- a/src/c_wrapper/memory_object.cpp +++ /dev/null @@ -1,116 +0,0 @@ -#include "memory_object.h" -#include "context.h" -#include "event.h" -#include "command_queue.h" -#include "clhelper.h" - -template class clobj; -template void print_arg(std::ostream&, const cl_mem&, bool); -template void print_buf(std::ostream&, const cl_mem*, - size_t, ArgType, bool, bool); - -generic_info -memory_object::get_info(cl_uint param_name) const -{ - switch ((cl_mem_info)param_name) { - case CL_MEM_TYPE: - return pyopencl_get_int_info(cl_mem_object_type, MemObject, - PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_MEM_FLAGS: - return pyopencl_get_int_info(cl_mem_flags, MemObject, - PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_MEM_SIZE: - return pyopencl_get_int_info(size_t, MemObject, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_MEM_HOST_PTR: - throw clerror("MemoryObject.get_info", CL_INVALID_VALUE, - "Use MemoryObject.get_host_array to get " - "host pointer."); - case CL_MEM_MAP_COUNT: - case CL_MEM_REFERENCE_COUNT: - return pyopencl_get_int_info(cl_uint, MemObject, - PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_MEM_CONTEXT: - return pyopencl_get_opaque_info(context, MemObject, PYOPENCL_CL_CASTABLE_THIS, param_name); - -#if PYOPENCL_CL_VERSION >= 0x1010 - // TODO - // case CL_MEM_ASSOCIATED_MEMOBJECT: - // { - // cl_mem param_value; - // PYOPENCL_CALL_GUARDED(clGetMemObjectInfo, (this, param_name, sizeof(param_value), ¶m_value, 0)); - // if (param_value == 0) - // { - // // no associated memory object? no problem. - // return py::object(); - // } - - // return create_mem_object_wrapper(param_value); - // } - case CL_MEM_OFFSET: - return pyopencl_get_int_info(size_t, MemObject, PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - case CL_MEM_USES_SVM_POINTER: - return pyopencl_get_int_info(cl_bool, MemObject, PYOPENCL_CL_CASTABLE_THIS, param_name); -#endif - - default: - throw clerror("MemoryObject.get_info", CL_INVALID_VALUE); - } -} - -memory_object::~memory_object() -{ - if (!m_valid.exchange(false)) - return; - pyopencl_call_guarded_cleanup(clReleaseMemObject, PYOPENCL_CL_CASTABLE_THIS); -} - -// c wrapper - -// Memory Object -error* -memory_object__release(clobj_t obj) -{ - return c_handle_error([&] { - static_cast(obj)->release(); - }); -} - -error* -memory_object__get_host_array(clobj_t _obj, void **hostptr, size_t *size) -{ - auto obj = static_cast(_obj); - return c_handle_error([&] { - cl_mem_flags flags; - pyopencl_call_guarded(clGetMemObjectInfo, obj, CL_MEM_FLAGS, - size_arg(flags), nullptr); - if (!(flags & CL_MEM_USE_HOST_PTR)) - throw clerror("MemoryObject.get_host_array", CL_INVALID_VALUE, - "Only MemoryObject with USE_HOST_PTR " - "is supported."); - pyopencl_call_guarded(clGetMemObjectInfo, obj, CL_MEM_HOST_PTR, - size_arg(*hostptr), nullptr); - pyopencl_call_guarded(clGetMemObjectInfo, obj, CL_MEM_SIZE, - size_arg(*size), nullptr); - }); -} - -error* -enqueue_migrate_mem_objects(clobj_t *evt, clobj_t _queue, - const clobj_t *_mem_obj, uint32_t num_mem_obj, - cl_mem_migration_flags flags, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - const auto mem_obj = buf_from_class(_mem_obj, num_mem_obj); - auto queue = static_cast(_queue); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded(clEnqueueMigrateMemObjects, queue, - mem_obj, flags, wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clEnqueueMigrateMemObjects, "CL 1.2") -#endif -} diff --git a/src/c_wrapper/memory_object.h b/src/c_wrapper/memory_object.h deleted file mode 100644 index 635dc470..00000000 --- a/src/c_wrapper/memory_object.h +++ /dev/null @@ -1,56 +0,0 @@ -#include "error.h" -#include - -#ifndef __PYOPENCL_MEMORY_OBJECT_H -#define __PYOPENCL_MEMORY_OBJECT_H - -// {{{ memory_object - -extern template class clobj; -extern template void print_arg(std::ostream&, const cl_mem&, bool); -extern template void print_buf(std::ostream&, const cl_mem*, - size_t, ArgType, bool, bool); - -class memory_object : public clobj { -private: - mutable volatile std::atomic_bool m_valid; -public: - constexpr static const char *class_name = "MEMORY_OBJECT"; - PYOPENCL_INLINE - memory_object(cl_mem mem, bool retain) - : clobj(mem), m_valid(true) - { - if (retain) { - pyopencl_call_guarded(clRetainMemObject, PYOPENCL_CL_CASTABLE_THIS); - } - } - PYOPENCL_INLINE - memory_object(const memory_object &mem) - : memory_object(mem.data(), true) - {} - ~memory_object(); - generic_info get_info(cl_uint param_name) const; - void - release() const - { - if (PYOPENCL_UNLIKELY(!m_valid.exchange(false))) { - throw clerror("MemoryObject.release", CL_INVALID_VALUE, - "trying to double-unref mem object"); - } - pyopencl_call_guarded(clReleaseMemObject, PYOPENCL_CL_CASTABLE_THIS); - } -#if 0 - PYOPENCL_USE_RESULT size_t - size() const - { - size_t param_value; - pyopencl_call_guarded(clGetMemObjectInfo, this, CL_MEM_SIZE, - size_arg(param_value), nullptr); - return param_value; - } -#endif -}; - -// }}} - -#endif diff --git a/src/c_wrapper/mingw-std-threads b/src/c_wrapper/mingw-std-threads deleted file mode 160000 index 776ce7fa..00000000 --- a/src/c_wrapper/mingw-std-threads +++ /dev/null @@ -1 +0,0 @@ -Subproject commit 776ce7faf9368ec9588ee77458799c281cb25737 diff --git a/src/c_wrapper/platform.cpp b/src/c_wrapper/platform.cpp deleted file mode 100644 index 21a896b2..00000000 --- a/src/c_wrapper/platform.cpp +++ /dev/null @@ -1,109 +0,0 @@ -#include "platform.h" -#include "device.h" -#include "clhelper.h" - -#include - -template class clobj; -template void print_arg(std::ostream&, - const cl_platform_id&, bool); -template void print_clobj(std::ostream&, const platform*); -template void print_buf(std::ostream&, const cl_platform_id*, - size_t, ArgType, bool, bool); - -generic_info -platform::get_info(cl_uint param_name) const -{ - switch ((cl_platform_info)param_name) { - case CL_PLATFORM_PROFILE: - case CL_PLATFORM_VERSION: - case CL_PLATFORM_NAME: - case CL_PLATFORM_VENDOR: -#if !(defined(CL_PLATFORM_NVIDIA) && CL_PLATFORM_NVIDIA == 0x3001) - case CL_PLATFORM_EXTENSIONS: -#endif - return pyopencl_get_str_info(Platform, PYOPENCL_CL_CASTABLE_THIS, param_name); - default: - throw clerror("Platform.get_info", CL_INVALID_VALUE); - } -} - -void -platform::get_version(cl_platform_id plat, int *major, int *minor) -{ - char s_buff[128]; - size_t size; - pyopencl_buf d_buff(0); - char *name = s_buff; - pyopencl_call_guarded(clGetPlatformInfo, plat, CL_PLATFORM_VERSION, - 0, nullptr, buf_arg(size)); - if (PYOPENCL_UNLIKELY(size > sizeof(s_buff))) { - d_buff.resize(size); - name = d_buff.get(); - } - pyopencl_call_guarded(clGetPlatformInfo, plat, CL_PLATFORM_VERSION, - size_arg(name, size), buf_arg(size)); - *major = *minor = -1; - sscanf(name, "OpenCL %d.%d", major, minor); - // Well, hopefully there won't be a negative OpenCL version =) - if (*major < 0 || *minor < 0) { - throw clerror("Platform.get_version", CL_INVALID_VALUE, - "platform returned non-conformant " - "platform version string"); - } -} - -// c wrapper - -error* -get_platforms(clobj_t **_platforms, uint32_t *num_platforms) -{ - return c_handle_error([&] { - *num_platforms = 0; - pyopencl_call_guarded(clGetPlatformIDs, 0, nullptr, - buf_arg(*num_platforms)); - pyopencl_buf platforms(*num_platforms); - pyopencl_call_guarded(clGetPlatformIDs, platforms, - buf_arg(*num_platforms)); - *_platforms = buf_to_base(platforms).release(); - }); -} - -error* -platform__get_devices(clobj_t _plat, clobj_t **_devices, - uint32_t *num_devices, cl_device_type devtype) -{ - auto plat = static_cast(_plat); - return c_handle_error([&] { - *num_devices = 0; - try { - pyopencl_call_guarded(clGetDeviceIDs, plat, devtype, 0, nullptr, - buf_arg(*num_devices)); - } catch (const clerror &e) { - if (e.code() != CL_DEVICE_NOT_FOUND) - throw e; - *num_devices = 0; - } - if (*num_devices == 0) { - *_devices = nullptr; - return; - } - pyopencl_buf devices(*num_devices); - pyopencl_call_guarded(clGetDeviceIDs, plat, devtype, devices, - buf_arg(*num_devices)); - *_devices = buf_to_base(devices).release(); - }); -} - -error* -platform__unload_compiler(clobj_t plat) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - return c_handle_error([&] { - pyopencl_call_guarded(clUnloadPlatformCompiler, - static_cast(plat)); - }); -#else - PYOPENCL_UNSUPPORTED(clUnloadPlatformCompiler, "CL 1.1 and below") -#endif -} diff --git a/src/c_wrapper/platform.h b/src/c_wrapper/platform.h deleted file mode 100644 index 1bad5c29..00000000 --- a/src/c_wrapper/platform.h +++ /dev/null @@ -1,27 +0,0 @@ -#include "error.h" - -#ifndef __PYOPENCL_PLATFORM_H -#define __PYOPENCL_PLATFORM_H - -// {{{ platform - -extern template class clobj; -extern template void print_arg(std::ostream&, - const cl_platform_id&, bool); -extern template void print_buf( - std::ostream&, const cl_platform_id*, size_t, ArgType, bool, bool); - -class platform : public clobj { -public: - static void get_version(cl_platform_id plat, int *major, int *minor); - using clobj::clobj; - PYOPENCL_DEF_CL_CLASS(PLATFORM); - - generic_info get_info(cl_uint param_name) const; -}; - -extern template void print_clobj(std::ostream&, const platform*); - -// }}} - -#endif diff --git a/src/c_wrapper/program.cpp b/src/c_wrapper/program.cpp deleted file mode 100644 index a0535c06..00000000 --- a/src/c_wrapper/program.cpp +++ /dev/null @@ -1,269 +0,0 @@ -#include "program.h" -#include "device.h" -#include "context.h" -#include "clhelper.h" -#include "kernel.h" - -template class clobj; -template void print_arg(std::ostream&, const cl_program&, bool); -template void print_clobj(std::ostream&, const program*); -template void print_buf(std::ostream&, const cl_program*, - size_t, ArgType, bool, bool); - -PYOPENCL_USE_RESULT static PYOPENCL_INLINE program* -new_program(cl_program prog, program_kind_type progkind=KND_UNKNOWN) -{ - return pyopencl_convert_obj(program, clReleaseProgram, prog, progkind); -} - -program::~program() -{ - pyopencl_call_guarded_cleanup(clReleaseProgram, PYOPENCL_CL_CASTABLE_THIS); -} - -generic_info -program::get_info(cl_uint param) const -{ - switch ((cl_program_info)param) { - case CL_PROGRAM_CONTEXT: - return pyopencl_get_opaque_info(context, Program, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_PROGRAM_REFERENCE_COUNT: - case CL_PROGRAM_NUM_DEVICES: - return pyopencl_get_int_info(cl_uint, Program, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_PROGRAM_DEVICES: - return pyopencl_get_opaque_array_info(device, Program, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_PROGRAM_SOURCE: - return pyopencl_get_str_info(Program, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_PROGRAM_BINARY_SIZES: - return pyopencl_get_array_info(size_t, Program, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_PROGRAM_BINARIES: { - auto sizes = pyopencl_get_vec_info(size_t, Program, PYOPENCL_CL_CASTABLE_THIS, - CL_PROGRAM_BINARY_SIZES); - pyopencl_buf result_ptrs(sizes.len()); - for (size_t i = 0;i < sizes.len();i++) { - result_ptrs[i] = (char*)malloc(sizes[i]); - } - try { - pyopencl_call_guarded(clGetProgramInfo, PYOPENCL_CL_CASTABLE_THIS, CL_PROGRAM_BINARIES, - sizes.len() * sizeof(char*), - result_ptrs.get(), nullptr); - } catch (...) { - for (size_t i = 0;i < sizes.len();i++) { - free(result_ptrs[i]); - } - } - pyopencl_buf gis(sizes.len()); - for (size_t i = 0;i < sizes.len();i++) { - gis[i] = make_generic_info( - CLASS_NONE, - _copy_str(std::string("char[") + tostring(sizes[i]) + "]"), - true, - result_ptrs[i], - true); - } - return pyopencl_convert_array_info(generic_info, gis); - } - -#if PYOPENCL_CL_VERSION >= 0x1020 - case CL_PROGRAM_NUM_KERNELS: - return pyopencl_get_int_info(size_t, Program, PYOPENCL_CL_CASTABLE_THIS, param); - case CL_PROGRAM_KERNEL_NAMES: - return pyopencl_get_str_info(Program, PYOPENCL_CL_CASTABLE_THIS, param); -#endif - default: - throw clerror("Program.get_info", CL_INVALID_VALUE); - } -} - -generic_info -program::get_build_info(const device *dev, cl_program_build_info param) const -{ - switch (param) { - case CL_PROGRAM_BUILD_STATUS: - return pyopencl_get_int_info(cl_build_status, ProgramBuild, - PYOPENCL_CL_CASTABLE_THIS, dev, param); - case CL_PROGRAM_BUILD_OPTIONS: - case CL_PROGRAM_BUILD_LOG: - return pyopencl_get_str_info(ProgramBuild, PYOPENCL_CL_CASTABLE_THIS, dev, param); -#if PYOPENCL_CL_VERSION >= 0x1020 - case CL_PROGRAM_BINARY_TYPE: - return pyopencl_get_int_info(cl_program_binary_type, ProgramBuild, - PYOPENCL_CL_CASTABLE_THIS, dev, param); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - case CL_PROGRAM_BUILD_GLOBAL_VARIABLE_TOTAL_SIZE: - return pyopencl_get_int_info(size_t, ProgramBuild, - PYOPENCL_CL_CASTABLE_THIS, dev, param); -#endif - default: - throw clerror("Program.get_build_info", CL_INVALID_VALUE); - } -} - -#if PYOPENCL_CL_VERSION >= 0x1020 -void -program::compile(const char *opts, const clobj_t *_devs, size_t num_devs, - const clobj_t *_prgs, const char *const *names, - size_t num_hdrs) -{ - const auto devs = buf_from_class(_devs, num_devs); - const auto prgs = buf_from_class(_prgs, num_hdrs); - pyopencl_call_guarded(clCompileProgram, PYOPENCL_CL_CASTABLE_THIS, devs, opts, prgs, - buf_arg(names, num_hdrs), nullptr, nullptr); -} -#endif - -pyopencl_buf -program::all_kernels() -{ - cl_uint num_knls; - pyopencl_call_guarded(clCreateKernelsInProgram, PYOPENCL_CL_CASTABLE_THIS, 0, nullptr, - buf_arg(num_knls)); - pyopencl_buf knls(num_knls); - pyopencl_call_guarded(clCreateKernelsInProgram, PYOPENCL_CL_CASTABLE_THIS, knls, - buf_arg(num_knls)); - return buf_to_base(knls, true); -} - -// c wrapper - -// Program -error* -create_program_with_source(clobj_t *prog, clobj_t _ctx, const char *_src) -{ - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - const auto &src = _src; - const size_t length = strlen(src); - cl_program result = pyopencl_call_guarded( - clCreateProgramWithSource, ctx, len_arg(src), buf_arg(length)); - *prog = new_program(result, KND_SOURCE); - }); -} - -error* -create_program_with_il(clobj_t *prog, clobj_t _ctx, void *il, size_t length) -{ -#if PYOPENCL_CL_VERSION >= 0x2010 - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - cl_program result = pyopencl_call_guarded( - clCreateProgramWithIL, ctx, il, length); - *prog = new_program(result, KND_SOURCE); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clCreateProgramWithIL, "CL 2.1") -#endif -} - -error* -create_program_with_binary(clobj_t *prog, clobj_t _ctx, - cl_uint num_devices, const clobj_t *devices, - const unsigned char **binaries, size_t *binary_sizes) -{ - auto ctx = static_cast(_ctx); - const auto devs = buf_from_class(devices, num_devices); - pyopencl_buf binary_statuses(num_devices); - return c_handle_error([&] { - cl_program result = pyopencl_call_guarded( - clCreateProgramWithBinary, ctx, devs, - binary_sizes, binaries, binary_statuses.get()); - // for (cl_uint i = 0; i < num_devices; ++i) - // std::cout << i << ":" << binary_statuses[i] << std::endl; - *prog = new_program(result, KND_BINARY); - }); -} - -error* -program__build(clobj_t _prog, const char *options, - cl_uint num_devices, const clobj_t *_devices) -{ - auto prog = static_cast(_prog); - const auto devices = buf_from_class(_devices, num_devices); - return c_handle_error([&] { - pyopencl_call_guarded(clBuildProgram, prog, devices, options, - nullptr, nullptr); - }); -} - -error* -program__kind(clobj_t prog, int *kind) -{ - return c_handle_error([&] { - *kind = static_cast(prog)->kind(); - }); -} - -error* -program__get_build_info(clobj_t _prog, clobj_t _dev, - cl_program_build_info param, generic_info *out) -{ - auto prog = static_cast(_prog); - auto dev = static_cast(_dev); - return c_handle_error([&] { - *out = prog->get_build_info(dev, param); - }); -} - -error* -program__create_with_builtin_kernels(clobj_t *_prg, clobj_t _ctx, - const clobj_t *_devs, uint32_t num_devs, - const char *names) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - const auto devs = buf_from_class(_devs, num_devs); - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - auto prg = pyopencl_call_guarded(clCreateProgramWithBuiltInKernels, - ctx, devs, names); - *_prg = new_program(prg); - }); -#else - PYOPENCL_UNSUPPORTED(clCreateProgramWithBuiltInKernels, "CL 1.1 and below") -#endif -} - -error* -program__compile(clobj_t _prg, const char *opts, const clobj_t *_devs, - size_t num_devs, const clobj_t *_prgs, - const char *const *names, size_t num_hdrs) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - auto prg = static_cast(_prg); - return c_handle_error([&] { - prg->compile(opts, _devs, num_devs, _prgs, names, num_hdrs); - }); -#else - PYOPENCL_UNSUPPORTED(clCompileProgram, "CL 1.1 and below") -#endif -} - -error* -program__link(clobj_t *_prg, clobj_t _ctx, const clobj_t *_prgs, - size_t num_prgs, const char *opts, const clobj_t *_devs, - size_t num_devs) -{ -#if PYOPENCL_CL_VERSION >= 0x1020 - const auto devs = buf_from_class(_devs, num_devs); - const auto prgs = buf_from_class(_prgs, num_prgs); - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - auto prg = pyopencl_call_guarded(clLinkProgram, ctx, devs, opts, - prgs, nullptr, nullptr); - *_prg = new_program(prg); - }); -#else - PYOPENCL_UNSUPPORTED(clLinkProgram, "CL 1.1 and below") -#endif -} - -error* -program__all_kernels(clobj_t _prg, clobj_t **_knl, uint32_t *size) -{ - auto prg = static_cast(_prg); - return c_handle_error([&] { - auto knls = prg->all_kernels(); - *size = knls.len(); - *_knl = knls.release(); - }); -} diff --git a/src/c_wrapper/program.h b/src/c_wrapper/program.h deleted file mode 100644 index 63d2fc76..00000000 --- a/src/c_wrapper/program.h +++ /dev/null @@ -1,58 +0,0 @@ -#include "clhelper.h" - -#ifndef __PYOPENCL_PROGRAM_H -#define __PYOPENCL_PROGRAM_H - -class device; - -// {{{ program - -extern template class clobj; -extern template void print_arg(std::ostream&, - const cl_program&, bool); -extern template void print_buf(std::ostream&, const cl_program*, - size_t, ArgType, bool, bool); - -class program : public clobj { -private: - program_kind_type m_program_kind; - -public: - PYOPENCL_DEF_CL_CLASS(PROGRAM); - PYOPENCL_INLINE - program(cl_program prog, bool retain, - program_kind_type progkind=KND_UNKNOWN) - : clobj(prog), m_program_kind(progkind) - { - if (retain) { - pyopencl_call_guarded(clRetainProgram, PYOPENCL_CL_CASTABLE_THIS); - } - } - ~program(); - PYOPENCL_USE_RESULT PYOPENCL_INLINE program_kind_type - kind() const - { - return m_program_kind; - } - PYOPENCL_USE_RESULT pyopencl_buf - get_info__devices() const - { - return pyopencl_get_vec_info(cl_device_id, Program, PYOPENCL_CL_CASTABLE_THIS, - CL_PROGRAM_DEVICES); - } - generic_info get_info(cl_uint param_name) const; - PYOPENCL_USE_RESULT generic_info - get_build_info(const device *dev, cl_program_build_info param_name) const; -#if PYOPENCL_CL_VERSION >= 0x1020 - void compile(const char *opts, const clobj_t *_devs, size_t num_devs, - const clobj_t *_prgs, const char *const *names, - size_t num_hdrs); -#endif - pyopencl_buf all_kernels(); -}; - -extern template void print_clobj(std::ostream&, const program*); - -// }}} - -#endif diff --git a/src/c_wrapper/pyhelper.cpp b/src/c_wrapper/pyhelper.cpp deleted file mode 100644 index 7397d12b..00000000 --- a/src/c_wrapper/pyhelper.cpp +++ /dev/null @@ -1,18 +0,0 @@ -#include "pyhelper.h" - -namespace py { -WrapFunc gc; -WrapFunc ref; -WrapFunc deref; -WrapFunc call; -} - -void -set_py_funcs(int (*_gc)(), void *(*_ref)(void*), void (*_deref)(void*), - void (*_call)(void*, cl_int)) -{ - py::gc = _gc; - py::ref = _ref; - py::deref = _deref; - py::call = _call; -} diff --git a/src/c_wrapper/pyhelper.h b/src/c_wrapper/pyhelper.h deleted file mode 100644 index 50c08402..00000000 --- a/src/c_wrapper/pyhelper.h +++ /dev/null @@ -1,43 +0,0 @@ -#ifndef __PYOPENCL_PYHELPER_H -#define __PYOPENCL_PYHELPER_H - -#include "wrap_cl.h" -#include "function.h" - -template -class WrapFunc; - -template -class WrapFunc { - typedef Ret (*_FuncType)(Args...); - _FuncType m_func; - static PYOPENCL_INLINE _FuncType - check_func(_FuncType f) - { - return f ? f : ([] (Args...) {return Ret();}); - } -public: - WrapFunc(_FuncType func=nullptr) - : m_func(check_func(func)) - {} - Ret - operator()(Args... args) - { - return m_func(std::forward(args)...); - } - WrapFunc& - operator=(_FuncType func) - { - m_func = check_func(func); - return *this; - } -}; - -namespace py { -extern WrapFunc gc; -extern WrapFunc ref; -extern WrapFunc deref; -extern WrapFunc call; -} - -#endif diff --git a/src/c_wrapper/pyopencl_ext.h b/src/c_wrapper/pyopencl_ext.h deleted file mode 100644 index a9792d8b..00000000 --- a/src/c_wrapper/pyopencl_ext.h +++ /dev/null @@ -1,58 +0,0 @@ -#ifndef _PYOPENCL_EXT_H -#define _PYOPENCL_EXT_H - -#ifdef PYOPENCL_USE_SHIPPED_EXT - -#include "clinfo_ext.h" - -#else - -#if (defined(__APPLE__) && !defined(PYOPENCL_APPLE_USE_CL_H)) - -#include - -#else - -#include -#include - -#endif - -#ifndef CL_DEVICE_TOPOLOGY_TYPE_PCIE_AMD -#define CL_DEVICE_TOPOLOGY_TYPE_PCIE_AMD 1 - -typedef union -{ - struct { cl_uint type; cl_uint data[5]; } raw; - struct { cl_uint type; cl_char unused[17]; cl_char bus; cl_char device; cl_char function; } pcie; -} cl_device_topology_amd; -#endif - -/* {{{ these NV defines are often missing from the system headers */ - -#ifndef CL_DEVICE_KERNEL_EXEC_TIMEOUT_NV -#define CL_DEVICE_KERNEL_EXEC_TIMEOUT_NV 0x4005 -#endif -#ifndef CL_DEVICE_INTEGRATED_MEMORY_NV -#define CL_DEVICE_INTEGRATED_MEMORY_NV 0x4006 -#endif - -#ifndef CL_DEVICE_ATTRIBUTE_ASYNC_ENGINE_COUNT_NV -#define CL_DEVICE_ATTRIBUTE_ASYNC_ENGINE_COUNT_NV 0x4007 -#endif - -#ifndef CL_DEVICE_PCI_BUS_ID_NV -#define CL_DEVICE_PCI_BUS_ID_NV 0x4008 -#endif - -#ifndef CL_DEVICE_PCI_SLOT_ID_NV -#define CL_DEVICE_PCI_SLOT_ID_NV 0x4009 -#endif - -/* }}} */ - -#endif - -#endif - -/* vim: foldmethod=marker */ diff --git a/src/c_wrapper/sampler.cpp b/src/c_wrapper/sampler.cpp deleted file mode 100644 index b373c783..00000000 --- a/src/c_wrapper/sampler.cpp +++ /dev/null @@ -1,54 +0,0 @@ -#include "sampler.h" -#include "context.h" -#include "clhelper.h" - -template class clobj; -template void print_arg(std::ostream&, const cl_sampler&, bool); -template void print_clobj(std::ostream&, const sampler*); -template void print_buf(std::ostream&, const cl_sampler*, - size_t, ArgType, bool, bool); - -sampler::~sampler() -{ - pyopencl_call_guarded_cleanup(clReleaseSampler, PYOPENCL_CL_CASTABLE_THIS); -} - -generic_info -sampler::get_info(cl_uint param_name) const -{ - switch ((cl_sampler_info)param_name) { - case CL_SAMPLER_REFERENCE_COUNT: - return pyopencl_get_int_info(cl_uint, Sampler, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_SAMPLER_CONTEXT: - return pyopencl_get_opaque_info(context, Sampler, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_SAMPLER_ADDRESSING_MODE: - return pyopencl_get_int_info(cl_addressing_mode, Sampler, - PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_SAMPLER_FILTER_MODE: - return pyopencl_get_int_info(cl_filter_mode, Sampler, PYOPENCL_CL_CASTABLE_THIS, param_name); - case CL_SAMPLER_NORMALIZED_COORDS: - return pyopencl_get_int_info(cl_bool, Sampler, PYOPENCL_CL_CASTABLE_THIS, param_name); - -#if PYOPENCL_CL_VERSION >= 0x2000 - // TODO: MIP_FILTER_MODE, LOD_MIN, LOD_MAX -#endif - - default: - throw clerror("Sampler.get_info", CL_INVALID_VALUE); - } -} - -// c wrapper - -// Sampler -error* -create_sampler(clobj_t *samp, clobj_t _ctx, int norm_coords, - cl_addressing_mode am, cl_filter_mode fm) -{ - auto ctx = static_cast(_ctx); - return c_handle_error([&] { - *samp = new sampler(pyopencl_call_guarded(clCreateSampler, ctx, - norm_coords, am, fm), - false); - }); -} diff --git a/src/c_wrapper/sampler.h b/src/c_wrapper/sampler.h deleted file mode 100644 index 404b82e5..00000000 --- a/src/c_wrapper/sampler.h +++ /dev/null @@ -1,33 +0,0 @@ -#include "error.h" - -#ifndef __PYOPENCL_SAMPLER_H -#define __PYOPENCL_SAMPLER_H - -// {{{ sampler - -extern template class clobj; -extern template void print_arg(std::ostream&, - const cl_sampler&, bool); -extern template void print_buf(std::ostream&, const cl_sampler*, - size_t, ArgType, bool, bool); - -class sampler : public clobj { -public: - PYOPENCL_DEF_CL_CLASS(SAMPLER); - PYOPENCL_INLINE - sampler(cl_sampler samp, bool retain) - : clobj(samp) - { - if (retain) { - pyopencl_call_guarded(clRetainSampler, PYOPENCL_CL_CASTABLE_THIS); - } - } - ~sampler(); - generic_info get_info(cl_uint param_name) const; -}; - -extern template void print_clobj(std::ostream&, const sampler*); - -// }}} - -#endif diff --git a/src/c_wrapper/svm.cpp b/src/c_wrapper/svm.cpp deleted file mode 100644 index 8452ec99..00000000 --- a/src/c_wrapper/svm.cpp +++ /dev/null @@ -1,173 +0,0 @@ -#include "context.h" -#include "command_queue.h" -#include "event.h" - -error* -svm_alloc( - clobj_t _ctx, cl_mem_flags flags, size_t size, cl_uint alignment, - void **result) -{ -#if PYOPENCL_CL_VERSION >= 0x2000 - auto ctx = static_cast(_ctx); - return c_handle_retry_mem_error([&] { - *result = clSVMAlloc(ctx->data(), flags, size, alignment); - if (!*result) - throw clerror("clSVMalloc", CL_INVALID_VALUE, - "(allocation failure, unspecified reason)"); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clSVMAlloc, "CL 2.0") -#endif -} - - -error* -svm_free(clobj_t _ctx, void *svm_pointer) -{ -#if PYOPENCL_CL_VERSION >= 0x2000 - auto ctx = static_cast(_ctx); - // no error returns (?!) - clSVMFree(ctx->data(), svm_pointer); - return nullptr; -#else - PYOPENCL_UNSUPPORTED_BEFORE(clSVMFree, "CL 2.0") -#endif -} - - -error* -enqueue_svm_free( - clobj_t *evt, clobj_t _queue, - cl_uint num_svm_pointers, - void *svm_pointers[], - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x2000 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueSVMFree, queue, - num_svm_pointers, svm_pointers, - /* pfn_free_func*/ nullptr, - /* user_data */ nullptr, - wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clEnqueueSVMFree, "CL 2.0") -#endif -} - - -error* -enqueue_svm_memcpy( - clobj_t *evt, clobj_t _queue, - cl_bool is_blocking, - void *dst_ptr, const void *src_ptr, size_t size, - const clobj_t *_wait_for, uint32_t num_wait_for, void *pyobj) -{ -#if PYOPENCL_CL_VERSION >= 0x2000 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueSVMMemcpy, queue, - is_blocking, - dst_ptr, src_ptr, size, - wait_for, nanny_event_out(evt, pyobj)); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clEnqueueSVMMemcpy, "CL 2.0") -#endif -} - - -error* -enqueue_svm_memfill( - clobj_t *evt, clobj_t _queue, - void *svm_ptr, - const void *pattern, size_t pattern_size, size_t size, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x2000 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueSVMMemFill, queue, - svm_ptr, - pattern, pattern_size, size, - wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clEnqueueSVMMemFill, "CL 2.0") -#endif -} - - -error* -enqueue_svm_map( - clobj_t *evt, clobj_t _queue, - cl_bool blocking_map, cl_map_flags map_flags, - void *svm_ptr, size_t size, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x2000 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueSVMMap, queue, - blocking_map, map_flags, - svm_ptr, size, - wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clEnqueueSVMMap, "CL 2.0") -#endif -} - - -error* -enqueue_svm_unmap( - clobj_t *evt, clobj_t _queue, - void *svm_ptr, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x2000 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueSVMUnmap, queue, - svm_ptr, - wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clEnqueueSVMUnmap, "CL 2.0") -#endif -} - - -error* -enqueue_svm_migrate_mem( - clobj_t *evt, clobj_t _queue, - cl_uint num_svm_pointers, - const void **svm_pointers, - const size_t *sizes, - cl_mem_migration_flags flags, - const clobj_t *_wait_for, uint32_t num_wait_for) -{ -#if PYOPENCL_CL_VERSION >= 0x2010 - const auto wait_for = buf_from_class(_wait_for, num_wait_for); - auto queue = static_cast(_queue); - return c_handle_retry_mem_error([&] { - pyopencl_call_guarded( - clEnqueueSVMMigrateMem, queue, - num_svm_pointers, svm_pointers, sizes, flags, - wait_for, event_out(evt)); - }); -#else - PYOPENCL_UNSUPPORTED_BEFORE(clEnqueueSVMMigrateMem, "CL 2.1") -#endif -} diff --git a/src/c_wrapper/svm.h b/src/c_wrapper/svm.h deleted file mode 100644 index c0e39ec4..00000000 --- a/src/c_wrapper/svm.h +++ /dev/null @@ -1,4 +0,0 @@ -#ifndef __PYOPENCL_SVM_H -#define __PYOPENCL_SVM_H - -#endif diff --git a/src/c_wrapper/utils.h b/src/c_wrapper/utils.h deleted file mode 100644 index d1bbb7d0..00000000 --- a/src/c_wrapper/utils.h +++ /dev/null @@ -1,551 +0,0 @@ -#include "wrap_cl.h" -#include "function.h" -#include "debug.h" - -#include -#include -#include -#include - -#ifndef __PYOPENCL_UTILS_H -#define __PYOPENCL_UTILS_H - -#if (defined(__GNUC__) && (__GNUC__ > 2)) -# define PYOPENCL_EXPECT(exp, var) __builtin_expect(exp, var) -#else -# define PYOPENCL_EXPECT(exp, var) (exp) -#endif - -#define PYOPENCL_LIKELY(x) PYOPENCL_EXPECT(bool(x), true) -#define PYOPENCL_UNLIKELY(x) PYOPENCL_EXPECT(bool(x), false) - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE std::string -tostring(const T& v) -{ - std::ostringstream ostr; - ostr << v; - return ostr.str(); -} - -template -struct CLGenericArgPrinter { - static PYOPENCL_INLINE void - print(std::ostream &stm, T &arg) - { - stm << arg; - } -}; - -PYOPENCL_USE_RESULT static PYOPENCL_INLINE void* -cl_memdup(const void *p, size_t size) -{ - void *res = malloc(size); - memcpy(res, p, size); - return res; -} - -template -PYOPENCL_USE_RESULT static PYOPENCL_INLINE T* -cl_memdup(const T *p) -{ - // Not supported by libstdc++ yet... - // static_assert(std::is_trivially_copy_constructible::value); - return static_cast(cl_memdup(static_cast(p), sizeof(T))); -} - -enum class ArgType { - None, - SizeOf, - Length, -}; - -template -struct type_size : std::integral_constant {}; -template -struct type_size, void>::value> > : - std::integral_constant {}; - -template -static PYOPENCL_INLINE void -_print_buf_content(std::ostream &stm, const T *p, size_t len) -{ - if (len > 1) { - stm << "["; - } - for (size_t i = 0;i < len;i++) { - CLGenericArgPrinter::print(stm, p[i]); - if (i != len - 1) { - stm << ", "; - } - } - if (len > 1) { - stm << "]"; - } -} - -template<> -PYOPENCL_INLINE void -_print_buf_content(std::ostream &stm, const char *p, size_t len) -{ - dbg_print_str(stm, p, len); -} - -template<> -PYOPENCL_INLINE void -_print_buf_content(std::ostream &stm, - const unsigned char *p, size_t len) -{ - dbg_print_bytes(stm, p, len); -} - -template<> -PYOPENCL_INLINE void -_print_buf_content(std::ostream &stm, const void *p, size_t len) -{ - dbg_print_bytes(stm, static_cast(p), len); -} - -template -void -print_buf(std::ostream &stm, const T *p, size_t len, - ArgType arg_type, bool content, bool out) -{ - const size_t ele_size = type_size::value; - if (out) { - stm << "*(" << (const void*)p << "): "; - if (p) { - _print_buf_content(stm, p, len); - } else { - stm << "NULL"; - } - } else { - bool need_quote = content || arg_type != ArgType::None; - if (content) { - if (p) { - _print_buf_content(stm, p, len); - stm << " "; - } else { - stm << "NULL "; - } - } - if (need_quote) { - stm << "<"; - } - switch (arg_type) { - case ArgType::SizeOf: - stm << ele_size * len << ", "; - break; - case ArgType::Length: - stm << len << ", "; - break; - default: - break; - } - stm << (const void*)p; - if (need_quote) { - stm << ">"; - } - } -} - -template -void -print_arg(std::ostream &stm, const T &v, bool out) -{ - if (!out) { - stm << (const void*)&v; - } else { - stm << "*(" << (const void*)&v << "): " << v; - } -} -extern template void print_buf(std::ostream&, const char*, size_t, - ArgType, bool, bool); -extern template void print_buf(std::ostream&, const cl_int*, size_t, - ArgType, bool, bool); -extern template void print_buf(std::ostream&, const cl_uint*, size_t, - ArgType, bool, bool); -extern template void print_buf(std::ostream&, const cl_long*, size_t, - ArgType, bool, bool); -extern template void print_buf(std::ostream&, const cl_ulong*, size_t, - ArgType, bool, bool); -extern template void print_buf(std::ostream&, - const cl_image_format*, size_t, - ArgType, bool, bool); - -template<> -struct CLGenericArgPrinter { - static PYOPENCL_INLINE void - print(std::ostream &stm, std::nullptr_t&) - { - stm << (void*)nullptr; - } -}; - -template -struct CLGenericArgPrinter< - T, enable_if_t >::value || - std::is_same >::value> > { - static PYOPENCL_INLINE void - print(std::ostream &stm, const char *str) - { - dbg_print_str(stm, str); - } -}; - -template -class CLArg { -private: - T &m_arg; -public: - CLArg(T &arg) noexcept - : m_arg(arg) - {} - CLArg(CLArg &&other) noexcept - : m_arg(other.m_arg) - {} - PYOPENCL_INLINE T& - convert() noexcept - { - return m_arg; - } - PYOPENCL_INLINE void - print(std::ostream &stm) - { - CLGenericArgPrinter::print(stm, m_arg); - } -}; - -template<> -class CLArg : public CLArg { - cl_bool m_arg; -public: - CLArg(bool arg) noexcept - : CLArg(m_arg), m_arg(arg ? CL_TRUE : CL_FALSE) - {} - CLArg(CLArg &&other) noexcept - : CLArg(bool(other.m_arg)) - {} - PYOPENCL_INLINE void - print(std::ostream &stm) - { - stm << (m_arg ? "true" : "false"); - } -}; - -template -class ArgBuffer { -private: - T *m_buf; - size_t m_len; -protected: - PYOPENCL_INLINE void - set(T *buf) noexcept - { - m_buf = buf; - } -public: - typedef T type; - constexpr static ArgType arg_type = AT; - ArgBuffer(T *buf, size_t l) noexcept - : m_buf(buf), m_len(l) - {} - ArgBuffer(ArgBuffer &&other) noexcept - : ArgBuffer(other.m_buf, other.m_len) - {} - PYOPENCL_INLINE rm_const_t* - get() const noexcept - { - return const_cast*>(m_buf); - } - template - PYOPENCL_INLINE T2& - operator[](int i) const - { - return m_buf[i]; - } - PYOPENCL_INLINE size_t - len() const noexcept - { - return m_len; - } -}; - -template -struct _ToArgBuffer { - static PYOPENCL_INLINE ArgBuffer, AT> - convert(T &buf) - { - return ArgBuffer, AT>(&buf, 1); - } -}; - -template -static PYOPENCL_INLINE auto -buf_arg(T &&buf) -> decltype(_ToArgBuffer::convert(std::forward(buf))) -{ - return _ToArgBuffer::convert(std::forward(buf)); -} - -template -static PYOPENCL_INLINE ArgBuffer -buf_arg(T *buf, size_t l) -{ - return ArgBuffer(buf, l); -} - -template -static PYOPENCL_INLINE auto -size_arg(T&&... buf) - -> decltype(buf_arg(std::forward(buf)...)) -{ - return buf_arg(std::forward(buf)...); -} - -template -static PYOPENCL_INLINE auto -len_arg(T&&... buf) - -> decltype(buf_arg(std::forward(buf)...)) -{ - return buf_arg(std::forward(buf)...); -} - -template -struct _ArgBufferConverter; - -template -struct _ArgBufferConverter > { - static PYOPENCL_INLINE auto - convert(Buff &buff) -> decltype(buff.get()) - { - return buff.get(); - } -}; - -template -struct _ArgBufferConverter > { - static PYOPENCL_INLINE auto - convert(Buff &buff) - -> decltype(std::make_tuple(type_size::value * - buff.len(), buff.get())) - { - return std::make_tuple(type_size::value * - buff.len(), buff.get()); - } -}; - -template -struct _ArgBufferConverter > { - static PYOPENCL_INLINE auto - convert(Buff &buff) -> decltype(std::make_tuple(buff.len(), buff.get())) - { - return std::make_tuple(buff.len(), buff.get()); - } -}; - -template -class CLArg, - Buff>::value> > { -private: - Buff &m_buff; -public: - constexpr static bool is_out = !std::is_const::value; - CLArg(Buff &buff) noexcept - : m_buff(buff) - {} - CLArg(CLArg &&other) noexcept - : m_buff(other.m_buff) - {} - PYOPENCL_INLINE auto - convert() const noexcept - -> decltype(_ArgBufferConverter::convert(m_buff)) - { - return _ArgBufferConverter::convert(m_buff); - } - PYOPENCL_INLINE void - print(std::ostream &stm, bool out=false) - { - print_buf(stm, m_buff.get(), m_buff.len(), - Buff::arg_type, out || !is_out, out); - } -}; - -template -class ConstBuffer : public ArgBuffer { -private: - T m_intern_buf[n]; - ConstBuffer(ConstBuffer&&) = delete; - ConstBuffer() = delete; -public: - ConstBuffer(const T *buf, size_t l, T content=0) - : ArgBuffer(buf, n) - { - if (l < n) { - memcpy(m_intern_buf, buf, type_size::value * l); - for (size_t i = l;i < n;i++) { - m_intern_buf[i] = content; - } - this->set(m_intern_buf); - } - } -}; - -struct OutArg { -}; - -template -class CLArg::value> > { -private: - bool m_converted; - bool m_need_cleanup; - T &m_arg; -public: - constexpr static bool is_out = true; - CLArg(T &arg) - : m_converted(false), m_need_cleanup(false), m_arg(arg) - { - } - CLArg(CLArg &&other) noexcept - : m_converted(other.m_converted), m_need_cleanup(other.m_need_cleanup), - m_arg(other.m_arg) - { - other.m_need_cleanup = false; - } - PYOPENCL_INLINE auto - convert() -> decltype(m_arg.get()) - { - return m_arg.get(); - } - PYOPENCL_INLINE void - finish(bool converted) noexcept - { - m_need_cleanup = !converted; - } - PYOPENCL_INLINE void - post() - { - m_arg.convert(); - m_converted = true; - } - ~CLArg() - { - if (m_need_cleanup) { - m_arg.cleanup(m_converted); - } - } - PYOPENCL_INLINE void - print(std::ostream &stm, bool out=false) - { - m_arg.print(stm, out); - } -}; - -template -struct _D { - void operator()(T *p) { - free((void*)p); - } -}; - -template -class pyopencl_buf : public std::unique_ptr > { - size_t m_len; -public: - PYOPENCL_INLINE - pyopencl_buf(size_t len=1) - : std::unique_ptr >((T*)(len ? malloc(sizeof(T) * (len + 1)) : - nullptr)), m_len(len) - { - if (len) { - memset((void*)this->get(), 0, (len + 1) * sizeof(T)); - } - } - PYOPENCL_INLINE size_t - len() const - { - return m_len; - } - PYOPENCL_INLINE T& - operator[](int i) - { - return this->get()[i]; - } - PYOPENCL_INLINE const T& - operator[](int i) const - { - return this->get()[i]; - } - PYOPENCL_INLINE void - resize(size_t len) - { - if (len == m_len) - return; - m_len = len; - this->reset((T*)realloc((void*)this->release(), - (len + 1) * sizeof(T))); - } -}; - -template -using pyopencl_buf_ele_t = typename rm_ref_t::element_type; - -template -struct is_pyopencl_buf : std::false_type {}; - -template -struct is_pyopencl_buf< - T, enable_if_t >, - rm_ref_t >::value> > : std::true_type {}; - -template -struct _ToArgBuffer::value && - std::is_const >::value> > { - static PYOPENCL_INLINE ArgBuffer, AT> - convert(T &&buf) - { - return ArgBuffer, AT>(buf.get(), buf.len()); - } -}; - -template -struct _ToArgBuffer::value && - !std::is_const >::value> > { - static PYOPENCL_INLINE ArgBuffer, AT> - convert(T &&buf) - { - return ArgBuffer, AT>(buf.get(), buf.len()); - } -}; - -template -using __pyopencl_buf_arg_type = - rm_ref_t()))>; - -template -class CLArg::value> > - : public CLArg<__pyopencl_buf_arg_type > { - typedef __pyopencl_buf_arg_type BufType; - BufType m_buff; -public: - PYOPENCL_INLINE - CLArg(Buff &buff) noexcept - : CLArg(m_buff), m_buff(len_arg(buff)) - {} - PYOPENCL_INLINE - CLArg(CLArg &&other) noexcept - : CLArg(m_buff), m_buff(std::move(other.m_buff)) - {} -}; - -// FIXME -PYOPENCL_USE_RESULT static PYOPENCL_INLINE char* -_copy_str(const std::string& str) -{ - return strdup(str.c_str()); -} - -#endif diff --git a/src/c_wrapper/wrap_cl.cpp b/src/c_wrapper/wrap_cl.cpp deleted file mode 100644 index 1e001eb4..00000000 --- a/src/c_wrapper/wrap_cl.cpp +++ /dev/null @@ -1,123 +0,0 @@ -#include "pyhelper.h" -#include "clhelper.h" -#include "platform.h" -#include "device.h" -#include "context.h" -#include "command_queue.h" -#include "event.h" -#include "memory_object.h" -#include "image.h" -#include "gl_obj.h" -#include "memory_map.h" -#include "buffer.h" -#include "sampler.h" -#include "program.h" -#include "kernel.h" - -template void print_buf(std::ostream&, const char*, size_t, - ArgType, bool, bool); -template void print_buf(std::ostream&, const cl_int*, size_t, - ArgType, bool, bool); -template void print_buf(std::ostream&, const cl_uint*, size_t, - ArgType, bool, bool); -template void print_buf(std::ostream&, const cl_long*, size_t, - ArgType, bool, bool); -template void print_buf(std::ostream&, const cl_ulong*, size_t, - ArgType, bool, bool); -template void print_buf(std::ostream&, - const cl_image_format*, size_t, - ArgType, bool, bool); - -// {{{ c wrapper - -// Generic functions -int -get_cl_version() -{ - return PYOPENCL_CL_VERSION; -} - -void -free_pointer(void *p) -{ - free(p); -} - -void -free_pointer_array(void **p, uint32_t size) -{ - for (uint32_t i = 0;i < size;i++) { - free(p[i]); - } -} - - -intptr_t -clobj__int_ptr(clobj_t obj) -{ - return PYOPENCL_LIKELY(obj) ? obj->intptr() : 0l; -} - -static PYOPENCL_INLINE clobj_t -_from_int_ptr(intptr_t ptr, class_t class_, bool retain) -{ - switch(class_) { - case CLASS_PLATFORM: - return clobj_from_int_ptr(ptr, retain); - case CLASS_DEVICE: - return clobj_from_int_ptr(ptr, retain); - case CLASS_KERNEL: - return clobj_from_int_ptr(ptr, retain); - case CLASS_CONTEXT: - return clobj_from_int_ptr(ptr, retain); - case CLASS_COMMAND_QUEUE: - return clobj_from_int_ptr(ptr, retain); - case CLASS_BUFFER: - return clobj_from_int_ptr(ptr, retain); - case CLASS_PROGRAM: - return clobj_from_int_ptr(ptr, retain); - case CLASS_EVENT: - return clobj_from_int_ptr(ptr, retain); - case CLASS_IMAGE: - return clobj_from_int_ptr(ptr, retain); - case CLASS_SAMPLER: - return clobj_from_int_ptr(ptr, retain); -#ifdef HAVE_GL - case CLASS_GL_BUFFER: - return clobj_from_int_ptr(ptr, retain); - case CLASS_GL_RENDERBUFFER: - return clobj_from_int_ptr(ptr, retain); -#endif - default: - throw clerror("unknown class", CL_INVALID_VALUE); - } -} - -error* -clobj__from_int_ptr(clobj_t *out, intptr_t ptr, class_t class_, int retain) -{ - return c_handle_error([&] { - *out = _from_int_ptr(ptr, class_, retain); - }); -} - -error* -clobj__get_info(clobj_t obj, cl_uint param, generic_info *out) -{ - return c_handle_error([&] { - if (PYOPENCL_UNLIKELY(!obj)) { - throw clerror("NULL input", CL_INVALID_VALUE); - } - *out = obj->get_info(param); - }); -} - -void -clobj__delete(clobj_t obj) -{ - delete obj; -} - -// }}} - -// vim: foldmethod=marker diff --git a/src/c_wrapper/wrap_cl.h b/src/c_wrapper/wrap_cl.h deleted file mode 100644 index 21ff9c08..00000000 --- a/src/c_wrapper/wrap_cl.h +++ /dev/null @@ -1,171 +0,0 @@ -#ifndef _WRAP_CL_H -#define _WRAP_CL_H - - -// CL 1.2 undecided: -// clSetPrintfCallback - -// {{{ includes - -#include - -#include "pyopencl_ext.h" - -#define CL_USE_DEPRECATED_OPENCL_1_1_APIS - -#if (defined(__APPLE__) && !defined(PYOPENCL_APPLE_USE_CL_H)) - -// {{{ Mac - -#define PYOPENCL_HAVE_EVENT_SET_CALLBACK - -#ifdef HAVE_GL - -#define PYOPENCL_GL_SHARING_VERSION 1 - -#include -#include -#include -#endif -// }}} - -#else - -// {{{ elsewhere - -#if defined(_WIN32) - -// {{{ Windows - -#define NOMINMAX -#include -#define strdup _strdup -#define strcasecmp _stricmp - -#if _MSC_VER >= 1900 /* VS 2015 and higher */ -#define PYOPENCL_HAVE_EVENT_SET_CALLBACK -#endif - -// }}} - -#else - -// {{{ non-Windows - -#include -#define PYOPENCL_HAVE_EVENT_SET_CALLBACK - -// }}} - -#endif - -#ifdef HAVE_GL -#include -#include -#endif - -#if defined(cl_khr_gl_sharing) && (cl_khr_gl_sharing >= 1) -#define PYOPENCL_GL_SHARING_VERSION cl_khr_gl_sharing -#endif - -// }}} - -#endif - -// }}} - - -// {{{ version handling - -#ifdef PYOPENCL_PRETEND_CL_VERSION -#define PYOPENCL_CL_VERSION PYOPENCL_PRETEND_CL_VERSION -#else - -#if defined(CL_VERSION_2_2) -#define PYOPENCL_CL_VERSION 0x2020 -#elif defined(CL_VERSION_2_1) -#define PYOPENCL_CL_VERSION 0x2010 -#elif defined(CL_VERSION_2_0) -#define PYOPENCL_CL_VERSION 0x2000 -#elif defined(CL_VERSION_1_2) -#define PYOPENCL_CL_VERSION 0x1020 -#elif defined(CL_VERSION_1_1) -#define PYOPENCL_CL_VERSION 0x1010 -#else -#define PYOPENCL_CL_VERSION 0x1000 -#endif - -#endif - -// }}} - -#ifndef CL_VERSION_2_0 -typedef void* CLeglImageKHR; -typedef void* CLeglDisplayKHR; -typedef void* CLeglSyncKHR; -typedef intptr_t cl_egl_image_properties_khr; -typedef cl_bitfield cl_device_svm_capabilities; -typedef cl_bitfield cl_svm_mem_flags; -typedef intptr_t cl_pipe_properties; -typedef cl_uint cl_pipe_info; -typedef cl_bitfield cl_sampler_properties; -typedef cl_uint cl_kernel_exec_info; -#endif - -#ifndef CL_VERSION_1_2 -typedef intptr_t cl_device_partition_property; -typedef cl_uint cl_kernel_arg_info; - -typedef struct _cl_image_desc { - cl_mem_object_type image_type; - size_t image_width; - size_t image_height; - size_t image_depth; - size_t image_array_size; - size_t image_row_pitch; - size_t image_slice_pitch; - cl_uint num_mip_levels; - cl_uint num_samples; - cl_mem buffer; -} cl_image_desc; - -typedef cl_bitfield cl_mem_migration_flags; -#endif - -#ifndef CL_VERSION_1_1 -typedef struct _cl_buffer_region { - size_t origin; - size_t size; -} cl_buffer_region; -#endif - -#ifndef cl_ext_migrate_memobject -typedef cl_bitfield cl_mem_migration_flags_ext; -#endif - -struct clbase; -typedef clbase *clobj_t; - -#ifdef __cplusplus -extern "C" { -#endif - -#include "wrap_cl_core.h" - -#ifdef HAVE_GL -#include "wrap_cl_gl_core.h" -#endif - -#ifdef __cplusplus -} -#endif - -#if defined __GNUC__ || defined __GNUG__ -#define PYOPENCL_USE_RESULT __attribute__((warn_unused_result)) -#else -#define PYOPENCL_USE_RESULT -#endif - -#endif - -// vim: foldmethod=marker diff --git a/src/c_wrapper/wrap_cl_core.h b/src/c_wrapper/wrap_cl_core.h deleted file mode 100644 index 184cd001..00000000 --- a/src/c_wrapper/wrap_cl_core.h +++ /dev/null @@ -1,399 +0,0 @@ -// Interface between C and Python - -struct clbase; -typedef struct clbase *clobj_t; - -// {{{ types - -typedef enum { - TYPE_FLOAT, - TYPE_INT, - TYPE_UINT, -} type_t; - -typedef enum { - KND_UNKNOWN, - KND_SOURCE, - KND_BINARY -} program_kind_type; - -typedef struct { - const char *routine; - const char *msg; - cl_int code; - int other; -} error; - -typedef enum { - CLASS_NONE, - CLASS_PLATFORM, - CLASS_DEVICE, - CLASS_KERNEL, - CLASS_CONTEXT, - CLASS_BUFFER, - CLASS_PROGRAM, - CLASS_EVENT, - CLASS_COMMAND_QUEUE, - CLASS_GL_BUFFER, - CLASS_GL_RENDERBUFFER, - CLASS_IMAGE, - CLASS_SAMPLER -} class_t; - -typedef struct { - class_t opaque_class; - const char *type; - bool free_type; - void *value; - bool free_value; -} generic_info; - -// }}} - -// {{{ generic functions - -int get_cl_version(); -void free_pointer(void*); -void free_pointer_array(void**, uint32_t size); -void set_py_funcs(int (*_gc)(), void *(*_ref)(void*), void (*_deref)(void*), - void (*_call)(void*, cl_int)); -int have_gl(); - -unsigned bitlog2(unsigned long v); -void populate_constants(void(*add)(const char*, const char*, int64_t value)); -int get_debug(); -void set_debug(int debug); - -// }}} - -// {{{ platform - -error *get_platforms(clobj_t **ptr_platforms, uint32_t *num_platforms); -error *platform__get_devices(clobj_t platform, clobj_t **ptr_devices, - uint32_t *num_devices, cl_device_type devtype); -error *platform__unload_compiler(clobj_t plat); - -// }}} - -// {{{ device -error *device__create_sub_devices(clobj_t _dev, clobj_t **_devs, - uint32_t *num_devices, - const cl_device_partition_property *props); - -// }}} - -// {{{ context - -error *create_context(clobj_t *ctx, const cl_context_properties *props, - cl_uint num_devices, const clobj_t *ptr_devices); -error *create_context_from_type(clobj_t *_ctx, - const cl_context_properties *props, - cl_device_type dev_type); -error *context__get_supported_image_formats(clobj_t context, cl_mem_flags flags, - cl_mem_object_type image_type, - generic_info *out); - -// }}} - -// {{{ command Queue - -error *create_command_queue(clobj_t *queue, clobj_t context, clobj_t device, - cl_command_queue_properties properties); -error *command_queue__finish(clobj_t queue); -error *command_queue__flush(clobj_t queue); - -// }}} - -// {{{ buffer -error *create_buffer(clobj_t *buffer, clobj_t context, cl_mem_flags flags, - size_t size, void *hostbuf); -error *buffer__get_sub_region(clobj_t *_sub_buf, clobj_t _buf, size_t orig, - size_t size, cl_mem_flags flags); - -// }}} - -// {{{ memory object - -error *memory_object__release(clobj_t obj); -error *memory_object__get_host_array(clobj_t, void **hostptr, size_t *size); - -// }}} - -// {{{ memory map - -error *memory_map__release(clobj_t _map, clobj_t _queue, - const clobj_t *_wait_for, uint32_t num_wait_for, - clobj_t *evt); -void *memory_map__data(clobj_t _map); - -// }}} - -// {{{ svm - -error* svm_alloc( - clobj_t _ctx, cl_mem_flags flags, size_t size, cl_uint alignment, - void **result); -error* svm_free(clobj_t _ctx, void *svm_pointer); -error* enqueue_svm_free( - clobj_t *evt, clobj_t _queue, - cl_uint num_svm_pointers, - void *svm_pointers[], - const clobj_t *_wait_for, uint32_t num_wait_for); -error* enqueue_svm_memcpy( - clobj_t *evt, clobj_t _queue, - cl_bool is_blocking, - void *dst_ptr, const void *src_ptr, size_t size, - const clobj_t *_wait_for, uint32_t num_wait_for, - void *pyobj); -error* enqueue_svm_memfill( - clobj_t *evt, clobj_t _queue, - void *svm_ptr, - const void *pattern, size_t pattern_size, size_t size, - const clobj_t *_wait_for, uint32_t num_wait_for); -error* enqueue_svm_map( - clobj_t *evt, clobj_t _queue, - cl_bool blocking_map, cl_map_flags map_flags, - void *svm_ptr, size_t size, - const clobj_t *_wait_for, uint32_t num_wait_for); -error* enqueue_svm_unmap( - clobj_t *evt, clobj_t _queue, - void *svm_ptr, - const clobj_t *_wait_for, uint32_t num_wait_for); -error* enqueue_svm_migrate_mem( - clobj_t *evt, clobj_t _queue, - cl_uint num_svm_pointers, - const void **svm_pointers, - const size_t *sizes, - cl_mem_migration_flags flags, - const clobj_t *_wait_for, uint32_t num_wait_for); - -// }}} - -// {{{ program - -error *create_program_with_source(clobj_t *program, clobj_t context, - const char *src); -error* create_program_with_il(clobj_t *prog, clobj_t _ctx, void *il, size_t length); -error *create_program_with_binary(clobj_t *program, clobj_t context, - cl_uint num_devices, const clobj_t *devices, - const unsigned char **binaries, - size_t *binary_sizes); -error *program__build(clobj_t program, const char *options, - cl_uint num_devices, const clobj_t *devices); -error *program__kind(clobj_t program, int *kind); -error *program__get_build_info(clobj_t program, clobj_t device, - cl_program_build_info param, generic_info *out); -error *program__create_with_builtin_kernels(clobj_t *_prg, clobj_t _ctx, - const clobj_t *_devs, - uint32_t num_devs, - const char *names); -error *program__compile(clobj_t _prg, const char *opts, const clobj_t *_devs, - size_t num_devs, const clobj_t *_prgs, - const char *const *names, size_t num_hdrs); -error *program__link(clobj_t *_prg, clobj_t _ctx, const clobj_t *_prgs, - size_t num_prgs, const char *opts, - const clobj_t *_devs, size_t num_devs); -error *program__all_kernels(clobj_t _prg, clobj_t **_knl, uint32_t *size); - -// }}} - -// {{{ sampler - -error *create_sampler(clobj_t *sampler, clobj_t context, int norm_coords, - cl_addressing_mode am, cl_filter_mode fm); - -// }}} - -// {{{ kernel - -error *create_kernel(clobj_t *kernel, clobj_t program, const char *name); -error *kernel__set_arg_null(clobj_t kernel, cl_uint arg_index); -error *kernel__set_arg_mem(clobj_t kernel, cl_uint arg_index, clobj_t mem); -error *kernel__set_arg_sampler(clobj_t kernel, cl_uint arg_index, - clobj_t sampler); -error *kernel__set_arg_buf(clobj_t kernel, cl_uint arg_index, - const void *buffer, size_t size); -error *kernel__set_arg_svm_pointer(clobj_t kernel, cl_uint arg_index, void *value); -error *kernel__get_work_group_info(clobj_t kernel, - cl_kernel_work_group_info param, - clobj_t device, generic_info *out); -error *kernel__get_arg_info(clobj_t _knl, cl_uint idx, - cl_kernel_arg_info param, generic_info *out); - -// }}} - -// {{{ image -error *create_image_2d(clobj_t *image, clobj_t context, cl_mem_flags flags, - cl_image_format *fmt, size_t width, size_t height, - size_t pitch, void *buffer); -error *create_image_3d(clobj_t *image, clobj_t context, cl_mem_flags flags, - cl_image_format *fmt, size_t width, size_t height, - size_t depth, size_t pitch_x, size_t pitch_y, - void *buffer); -error *create_image_from_desc(clobj_t *img, clobj_t _ctx, cl_mem_flags flags, - cl_image_format *fmt, cl_image_desc *desc, - void *buffer); -error *image__get_image_info(clobj_t img, cl_image_info param, - generic_info *out); -type_t image__get_fill_type(clobj_t img); -// }}} - -// {{{ event - -error *event__get_profiling_info(clobj_t event, cl_profiling_info param, - generic_info *out); -error *event__wait(clobj_t event); -error *event__set_callback(clobj_t _evt, cl_int type, void *pyobj); -error *wait_for_events(const clobj_t *_wait_for, uint32_t num_wait_for); - -// }}} - -// {{{ nanny event - -void *nanny_event__get_ward(clobj_t evt); - -// }}} - -// {{{ user event - -error *create_user_event(clobj_t *_evt, clobj_t _ctx); -error *user_event__set_status(clobj_t _evt, cl_int status); - -// }}} - -// {{{ enqueue_* -error *enqueue_nd_range_kernel(clobj_t *event, clobj_t queue, - clobj_t kernel, cl_uint work_dim, - const size_t *global_work_offset, - const size_t *global_work_size, - const size_t *local_work_size, - const clobj_t *wait_for, uint32_t num_wait_for); -error *enqueue_task(clobj_t *_evt, clobj_t _queue, clobj_t _knl, - const clobj_t *_wait_for, uint32_t num_wait_for); - -error *enqueue_marker_with_wait_list(clobj_t *event, clobj_t queue, - const clobj_t *wait_for, - uint32_t num_wait_for); -error *enqueue_barrier_with_wait_list(clobj_t *event, clobj_t queue, - const clobj_t *wait_for, - uint32_t num_wait_for); -error *enqueue_wait_for_events(clobj_t _queue, const clobj_t *_wait_for, - uint32_t num_wait_for); -error *enqueue_marker(clobj_t *event, clobj_t queue); -error *enqueue_barrier(clobj_t queue); -error *enqueue_migrate_mem_objects(clobj_t *evt, clobj_t _queue, - const clobj_t *_mem_obj, uint32_t, - cl_mem_migration_flags flags, - const clobj_t *_wait_for, uint32_t num_wait_for); - -// }}} - -// {{{ enqueue_*_buffer* - -error *enqueue_read_buffer(clobj_t *event, clobj_t queue, clobj_t mem, - void *buffer, size_t size, size_t device_offset, - const clobj_t *wait_for, uint32_t num_wait_for, - int is_blocking, void *pyobj); -error *enqueue_copy_buffer(clobj_t *event, clobj_t queue, clobj_t src, - clobj_t dst, ptrdiff_t byte_count, - size_t src_offset, size_t dst_offset, - const clobj_t *wait_for, uint32_t num_wait_for); -error *enqueue_write_buffer(clobj_t *event, clobj_t queue, clobj_t mem, - const void *buffer, size_t size, - size_t device_offset, const clobj_t *wait_for, - uint32_t num_wait_for, int is_blocking, - void *pyobj); -error *enqueue_map_buffer(clobj_t *_evt, clobj_t *mpa, clobj_t _queue, - clobj_t _mem, cl_map_flags flags, size_t offset, - size_t size, const clobj_t *_wait_for, - uint32_t num_wait_for, int block); -error *enqueue_fill_buffer(clobj_t *_evt, clobj_t _queue, clobj_t _mem, - void *pattern, size_t psize, size_t offset, - size_t size, const clobj_t *_wait_for, - uint32_t num_wait_for); -error *enqueue_read_buffer_rect(clobj_t *evt, clobj_t _queue, clobj_t _mem, - void *buf, const size_t *_buf_orig, - size_t buf_orig_l, const size_t *_host_orig, - size_t host_orig_l, const size_t *_reg, - size_t reg_l, const size_t *_buf_pitches, - size_t buf_pitches_l, - const size_t *_host_pitches, - size_t host_pitches_l, const clobj_t *_wait_for, - uint32_t num_wait_for, int block, void *pyobj); -error *enqueue_write_buffer_rect(clobj_t *evt, clobj_t _queue, clobj_t _mem, - void *buf, const size_t *_buf_orig, - size_t buf_orig_l, const size_t *_host_orig, - size_t host_orig_l, const size_t *_reg, - size_t reg_l, const size_t *_buf_pitches, - size_t buf_pitches_l, - const size_t *_host_pitches, - size_t host_pitches_l, - const clobj_t *_wait_for, - uint32_t num_wait_for, int block, void *pyobj); -error *enqueue_copy_buffer_rect(clobj_t *evt, clobj_t _queue, clobj_t _src, - clobj_t _dst, const size_t *_src_orig, - size_t src_orig_l, const size_t *_dst_orig, - size_t dst_orig_l, const size_t *_reg, - size_t reg_l, const size_t *_src_pitches, - size_t src_pitches_l, - const size_t *_dst_pitches, - size_t dst_pitches_l, const clobj_t *_wait_for, - uint32_t num_wait_for); - -// }}} - -// {{{ enqueue_*_image* - -error *enqueue_read_image(clobj_t *event, clobj_t queue, clobj_t mem, - const size_t *origin, size_t origin_l, - const size_t *region, size_t region_l, - void *buffer, size_t row_pitch, size_t slice_pitch, - const clobj_t *wait_for, uint32_t num_wait_for, - int is_blocking, void *pyobj); -error *enqueue_copy_image(clobj_t *_evt, clobj_t _queue, clobj_t _src, - clobj_t _dst, const size_t *_src_origin, - size_t src_origin_l, const size_t *_dst_origin, - size_t dst_origin_l, const size_t *_region, - size_t region_l, const clobj_t *_wait_for, - uint32_t num_wait_for); -error *enqueue_write_image(clobj_t *_evt, clobj_t _queue, clobj_t _mem, - const size_t *origin, size_t origin_l, - const size_t *region, size_t region_l, - const void *buffer, size_t row_pitch, - size_t slice_pitch, const clobj_t *_wait_for, - uint32_t num_wait_for, int is_blocking, - void *pyobj); -error *enqueue_map_image(clobj_t *_evt, clobj_t *map, clobj_t _queue, - clobj_t _mem, cl_map_flags flags, - const size_t *_origin, size_t origin_l, - const size_t *_region, size_t region_l, - size_t *row_pitch, size_t *slice_pitch, - const clobj_t *_wait_for, uint32_t num_wait_for, - int block); -error *enqueue_fill_image(clobj_t *evt, clobj_t _queue, clobj_t mem, - const void *color, const size_t *_origin, - size_t origin_l, const size_t *_region, - size_t region_l, const clobj_t *_wait_for, - uint32_t num_wait_for); -error *enqueue_copy_image_to_buffer(clobj_t *evt, clobj_t _queue, clobj_t _src, - clobj_t _dst, const size_t *_orig, size_t, - const size_t *_reg, size_t, size_t offset, - const clobj_t *_wait_for, uint32_t); -error *enqueue_copy_buffer_to_image(clobj_t *evt, clobj_t _queue, clobj_t _src, - clobj_t _dst, size_t offset, - const size_t *_orig, size_t, - const size_t *_reg, size_t, - const clobj_t *_wait_for, uint32_t); - -// }}} - -// {{{ cl object - -intptr_t clobj__int_ptr(clobj_t obj); -error *clobj__get_info(clobj_t obj, cl_uint param, generic_info *out); -void clobj__delete(clobj_t obj); -error *clobj__from_int_ptr(clobj_t *out, intptr_t ptr, class_t, int); - -// }}} - -// vim: foldmethod=marker diff --git a/src/c_wrapper/wrap_cl_gl_core.h b/src/c_wrapper/wrap_cl_gl_core.h deleted file mode 100644 index 606d7c1d..00000000 --- a/src/c_wrapper/wrap_cl_gl_core.h +++ /dev/null @@ -1,18 +0,0 @@ -// Interface between C and Python for GL related functions - -error* create_from_gl_texture(clobj_t *ptr, clobj_t _ctx, cl_mem_flags flags, - GLenum texture_target, GLint miplevel, - GLuint texture); -error *create_from_gl_buffer(clobj_t *ptr, clobj_t context, - cl_mem_flags flags, GLuint bufobj); -error *create_from_gl_renderbuffer(clobj_t *ptr, clobj_t context, - cl_mem_flags flags, GLuint bufobj); -error *enqueue_acquire_gl_objects( - clobj_t *event, clobj_t queue, const clobj_t *mem_objects, - uint32_t num_mem_objects, const clobj_t *wait_for, uint32_t num_wait_for); -error *enqueue_release_gl_objects( - clobj_t *event, clobj_t queue, const clobj_t *mem_objects, - uint32_t num_mem_objects, const clobj_t *wait_for, uint32_t num_wait_for); -cl_context_properties get_apple_cgl_share_group(); -error *get_gl_object_info(clobj_t mem, cl_gl_object_type *otype, - GLuint *gl_name); diff --git a/src/c_wrapper/wrap_constants.cpp b/src/c_wrapper/wrap_constants.cpp deleted file mode 100644 index 08ed2ede..00000000 --- a/src/c_wrapper/wrap_constants.cpp +++ /dev/null @@ -1,827 +0,0 @@ -#include "wrap_cl.h" -#include - -#ifdef CONST -#undef CONST -#endif - -extern "C" -void populate_constants(void(*add)(const char*, const char*, int64_t value)) -{ -#define _ADD_ATTR(TYPE, PREFIX, NAME, SUFFIX, ...) \ - add(TYPE, #NAME, CL_##PREFIX##NAME##SUFFIX) -#define ADD_ATTR(TYPE, PREFIX, NAME, ...) \ - _ADD_ATTR(TYPE, PREFIX, NAME, __VA_ARGS__) - - // program_kind - add("program_kind", "UNKNOWN", KND_UNKNOWN); - add("program_kind", "SOURCE", KND_SOURCE); - add("program_kind", "BINARY", KND_BINARY); - - // status_code - ADD_ATTR("status_code", , SUCCESS); - ADD_ATTR("status_code", , DEVICE_NOT_FOUND); - ADD_ATTR("status_code", , DEVICE_NOT_AVAILABLE); -#if !(defined(CL_PLATFORM_NVIDIA) && CL_PLATFORM_NVIDIA == 0x3001) - ADD_ATTR("status_code", , COMPILER_NOT_AVAILABLE); -#endif - ADD_ATTR("status_code", , MEM_OBJECT_ALLOCATION_FAILURE); - ADD_ATTR("status_code", , OUT_OF_RESOURCES); - ADD_ATTR("status_code", , OUT_OF_HOST_MEMORY); - ADD_ATTR("status_code", , PROFILING_INFO_NOT_AVAILABLE); - ADD_ATTR("status_code", , MEM_COPY_OVERLAP); - ADD_ATTR("status_code", , IMAGE_FORMAT_MISMATCH); - ADD_ATTR("status_code", , IMAGE_FORMAT_NOT_SUPPORTED); - ADD_ATTR("status_code", , BUILD_PROGRAM_FAILURE); - ADD_ATTR("status_code", , MAP_FAILURE); - - ADD_ATTR("status_code", , INVALID_VALUE); - ADD_ATTR("status_code", , INVALID_DEVICE_TYPE); - ADD_ATTR("status_code", , INVALID_PLATFORM); - ADD_ATTR("status_code", , INVALID_DEVICE); - ADD_ATTR("status_code", , INVALID_CONTEXT); - ADD_ATTR("status_code", , INVALID_QUEUE_PROPERTIES); - ADD_ATTR("status_code", , INVALID_COMMAND_QUEUE); - ADD_ATTR("status_code", , INVALID_HOST_PTR); - ADD_ATTR("status_code", , INVALID_MEM_OBJECT); - ADD_ATTR("status_code", , INVALID_IMAGE_FORMAT_DESCRIPTOR); - ADD_ATTR("status_code", , INVALID_IMAGE_SIZE); - ADD_ATTR("status_code", , INVALID_SAMPLER); - ADD_ATTR("status_code", , INVALID_BINARY); - ADD_ATTR("status_code", , INVALID_BUILD_OPTIONS); - ADD_ATTR("status_code", , INVALID_PROGRAM); - ADD_ATTR("status_code", , INVALID_PROGRAM_EXECUTABLE); - ADD_ATTR("status_code", , INVALID_KERNEL_NAME); - ADD_ATTR("status_code", , INVALID_KERNEL_DEFINITION); - ADD_ATTR("status_code", , INVALID_KERNEL); - ADD_ATTR("status_code", , INVALID_ARG_INDEX); - ADD_ATTR("status_code", , INVALID_ARG_VALUE); - ADD_ATTR("status_code", , INVALID_ARG_SIZE); - ADD_ATTR("status_code", , INVALID_KERNEL_ARGS); - ADD_ATTR("status_code", , INVALID_WORK_DIMENSION); - ADD_ATTR("status_code", , INVALID_WORK_GROUP_SIZE); - ADD_ATTR("status_code", , INVALID_WORK_ITEM_SIZE); - ADD_ATTR("status_code", , INVALID_GLOBAL_OFFSET); - ADD_ATTR("status_code", , INVALID_EVENT_WAIT_LIST); - ADD_ATTR("status_code", , INVALID_EVENT); - ADD_ATTR("status_code", , INVALID_OPERATION); - ADD_ATTR("status_code", , INVALID_GL_OBJECT); - ADD_ATTR("status_code", , INVALID_BUFFER_SIZE); - ADD_ATTR("status_code", , INVALID_MIP_LEVEL); - -#if defined(cl_khr_icd) && (cl_khr_icd >= 1) - ADD_ATTR("status_code", , PLATFORM_NOT_FOUND_KHR); -#endif - -#if defined(cl_khr_gl_sharing) && (cl_khr_gl_sharing >= 1) - ADD_ATTR("status_code", , INVALID_GL_SHAREGROUP_REFERENCE_KHR); -#endif - -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("status_code", , MISALIGNED_SUB_BUFFER_OFFSET); - ADD_ATTR("status_code", , EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST); - ADD_ATTR("status_code", , INVALID_GLOBAL_WORK_SIZE); -#endif - -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("status_code", , COMPILE_PROGRAM_FAILURE); - ADD_ATTR("status_code", , LINKER_NOT_AVAILABLE); - ADD_ATTR("status_code", , LINK_PROGRAM_FAILURE); - ADD_ATTR("status_code", , DEVICE_PARTITION_FAILED); - ADD_ATTR("status_code", , KERNEL_ARG_INFO_NOT_AVAILABLE); - ADD_ATTR("status_code", , INVALID_IMAGE_DESCRIPTOR); - ADD_ATTR("status_code", , INVALID_COMPILER_OPTIONS); - ADD_ATTR("status_code", , INVALID_LINKER_OPTIONS); - ADD_ATTR("status_code", , INVALID_DEVICE_PARTITION_COUNT); -#endif - -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("status_code", , INVALID_PIPE_SIZE); - ADD_ATTR("status_code", , INVALID_DEVICE_QUEUE); -#endif - - // platform_info - ADD_ATTR("platform_info", PLATFORM_, PROFILE); - ADD_ATTR("platform_info", PLATFORM_, VERSION); - ADD_ATTR("platform_info", PLATFORM_, NAME); - ADD_ATTR("platform_info", PLATFORM_, VENDOR); -#if !(defined(CL_PLATFORM_NVIDIA) && CL_PLATFORM_NVIDIA == 0x3001) - ADD_ATTR("platform_info", PLATFORM_, EXTENSIONS); -#endif - - - // device_type - ADD_ATTR("device_type", DEVICE_TYPE_, DEFAULT); - ADD_ATTR("device_type", DEVICE_TYPE_, CPU); - ADD_ATTR("device_type", DEVICE_TYPE_, GPU); - ADD_ATTR("device_type", DEVICE_TYPE_, ACCELERATOR); -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("device_type", DEVICE_TYPE_, CUSTOM); -#endif - ADD_ATTR("device_type", DEVICE_TYPE_, ALL); - - - // device_info - ADD_ATTR("device_info", DEVICE_, TYPE); - ADD_ATTR("device_info", DEVICE_, VENDOR_ID); - ADD_ATTR("device_info", DEVICE_, MAX_COMPUTE_UNITS); - ADD_ATTR("device_info", DEVICE_, MAX_WORK_ITEM_DIMENSIONS); - ADD_ATTR("device_info", DEVICE_, MAX_WORK_GROUP_SIZE); - ADD_ATTR("device_info", DEVICE_, MAX_WORK_ITEM_SIZES); - ADD_ATTR("device_info", DEVICE_, PREFERRED_VECTOR_WIDTH_CHAR); - ADD_ATTR("device_info", DEVICE_, PREFERRED_VECTOR_WIDTH_SHORT); - ADD_ATTR("device_info", DEVICE_, PREFERRED_VECTOR_WIDTH_INT); - ADD_ATTR("device_info", DEVICE_, PREFERRED_VECTOR_WIDTH_LONG); - ADD_ATTR("device_info", DEVICE_, PREFERRED_VECTOR_WIDTH_FLOAT); - ADD_ATTR("device_info", DEVICE_, PREFERRED_VECTOR_WIDTH_DOUBLE); - ADD_ATTR("device_info", DEVICE_, MAX_CLOCK_FREQUENCY); - ADD_ATTR("device_info", DEVICE_, ADDRESS_BITS); - ADD_ATTR("device_info", DEVICE_, MAX_READ_IMAGE_ARGS); - ADD_ATTR("device_info", DEVICE_, MAX_WRITE_IMAGE_ARGS); - ADD_ATTR("device_info", DEVICE_, MAX_MEM_ALLOC_SIZE); - ADD_ATTR("device_info", DEVICE_, IMAGE2D_MAX_WIDTH); - ADD_ATTR("device_info", DEVICE_, IMAGE2D_MAX_HEIGHT); - ADD_ATTR("device_info", DEVICE_, IMAGE3D_MAX_WIDTH); - ADD_ATTR("device_info", DEVICE_, IMAGE3D_MAX_HEIGHT); - ADD_ATTR("device_info", DEVICE_, IMAGE3D_MAX_DEPTH); - ADD_ATTR("device_info", DEVICE_, IMAGE_SUPPORT); - ADD_ATTR("device_info", DEVICE_, MAX_PARAMETER_SIZE); - ADD_ATTR("device_info", DEVICE_, MAX_SAMPLERS); - ADD_ATTR("device_info", DEVICE_, MEM_BASE_ADDR_ALIGN); - ADD_ATTR("device_info", DEVICE_, MIN_DATA_TYPE_ALIGN_SIZE); - ADD_ATTR("device_info", DEVICE_, SINGLE_FP_CONFIG); -#ifdef CL_DEVICE_DOUBLE_FP_CONFIG - ADD_ATTR("device_info", DEVICE_, DOUBLE_FP_CONFIG); -#endif -#ifdef CL_DEVICE_HALF_FP_CONFIG - ADD_ATTR("device_info", DEVICE_, HALF_FP_CONFIG); -#endif - ADD_ATTR("device_info", DEVICE_, GLOBAL_MEM_CACHE_TYPE); - ADD_ATTR("device_info", DEVICE_, GLOBAL_MEM_CACHELINE_SIZE); - ADD_ATTR("device_info", DEVICE_, GLOBAL_MEM_CACHE_SIZE); - ADD_ATTR("device_info", DEVICE_, GLOBAL_MEM_SIZE); - ADD_ATTR("device_info", DEVICE_, MAX_CONSTANT_BUFFER_SIZE); - ADD_ATTR("device_info", DEVICE_, MAX_CONSTANT_ARGS); - ADD_ATTR("device_info", DEVICE_, LOCAL_MEM_TYPE); - ADD_ATTR("device_info", DEVICE_, LOCAL_MEM_SIZE); - ADD_ATTR("device_info", DEVICE_, ERROR_CORRECTION_SUPPORT); - ADD_ATTR("device_info", DEVICE_, PROFILING_TIMER_RESOLUTION); - ADD_ATTR("device_info", DEVICE_, ENDIAN_LITTLE); - ADD_ATTR("device_info", DEVICE_, AVAILABLE); - ADD_ATTR("device_info", DEVICE_, COMPILER_AVAILABLE); - ADD_ATTR("device_info", DEVICE_, EXECUTION_CAPABILITIES); - ADD_ATTR("device_info", DEVICE_, QUEUE_PROPERTIES); -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("device_info", DEVICE_, QUEUE_ON_HOST_PROPERTIES); -#endif - ADD_ATTR("device_info", DEVICE_, NAME); - ADD_ATTR("device_info", DEVICE_, VENDOR); - ADD_ATTR("device_info", , DRIVER_VERSION); - ADD_ATTR("device_info", DEVICE_, VERSION); - ADD_ATTR("device_info", DEVICE_, PROFILE); - ADD_ATTR("device_info", DEVICE_, EXTENSIONS); - ADD_ATTR("device_info", DEVICE_, PLATFORM); -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("device_info", DEVICE_, PREFERRED_VECTOR_WIDTH_HALF); - ADD_ATTR("device_info", DEVICE_, HOST_UNIFIED_MEMORY); // deprecated in 2.0 - ADD_ATTR("device_info", DEVICE_, NATIVE_VECTOR_WIDTH_CHAR); - ADD_ATTR("device_info", DEVICE_, NATIVE_VECTOR_WIDTH_SHORT); - ADD_ATTR("device_info", DEVICE_, NATIVE_VECTOR_WIDTH_INT); - ADD_ATTR("device_info", DEVICE_, NATIVE_VECTOR_WIDTH_LONG); - ADD_ATTR("device_info", DEVICE_, NATIVE_VECTOR_WIDTH_FLOAT); - ADD_ATTR("device_info", DEVICE_, NATIVE_VECTOR_WIDTH_DOUBLE); - ADD_ATTR("device_info", DEVICE_, NATIVE_VECTOR_WIDTH_HALF); - ADD_ATTR("device_info", DEVICE_, OPENCL_C_VERSION); -#endif -#ifdef CL_DEVICE_COMPUTE_CAPABILITY_MAJOR_NV - ADD_ATTR("device_info", DEVICE_, COMPUTE_CAPABILITY_MAJOR_NV); - ADD_ATTR("device_info", DEVICE_, COMPUTE_CAPABILITY_MINOR_NV); - ADD_ATTR("device_info", DEVICE_, REGISTERS_PER_BLOCK_NV); - ADD_ATTR("device_info", DEVICE_, WARP_SIZE_NV); - ADD_ATTR("device_info", DEVICE_, GPU_OVERLAP_NV); - ADD_ATTR("device_info", DEVICE_, KERNEL_EXEC_TIMEOUT_NV); - ADD_ATTR("device_info", DEVICE_, INTEGRATED_MEMORY_NV); - // Nvidia specific device attributes, not defined in Khronos CL/cl_ext.h -#ifdef CL_DEVICE_ATTRIBUTE_ASYNC_ENGINE_COUNT_NV - ADD_ATTR("device_info", DEVICE_, ATTRIBUTE_ASYNC_ENGINE_COUNT_NV); -#endif -#ifdef CL_DEVICE_PCI_BUS_ID_NV - ADD_ATTR("device_info", DEVICE_, PCI_BUS_ID_NV); -#endif -#ifdef CL_DEVICE_PCI_SLOT_ID_NV - ADD_ATTR("device_info", DEVICE_, PCI_SLOT_ID_NV); -#endif -#endif -#ifdef CL_DEVICE_PROFILING_TIMER_OFFSET_AMD - ADD_ATTR("device_info", DEVICE_, PROFILING_TIMER_OFFSET_AMD); -#endif -#ifdef CL_DEVICE_TOPOLOGY_AMD - ADD_ATTR("device_info", DEVICE_, TOPOLOGY_AMD); -#endif -#ifdef CL_DEVICE_BOARD_NAME_AMD - ADD_ATTR("device_info", DEVICE_, BOARD_NAME_AMD); -#endif -#ifdef CL_DEVICE_GLOBAL_FREE_MEMORY_AMD - ADD_ATTR("device_info", DEVICE_, GLOBAL_FREE_MEMORY_AMD); -#endif -#ifdef CL_DEVICE_SIMD_PER_COMPUTE_UNIT_AMD - ADD_ATTR("device_info", DEVICE_, SIMD_PER_COMPUTE_UNIT_AMD); -#endif -#ifdef CL_DEVICE_SIMD_WIDTH_AMD - ADD_ATTR("device_info", DEVICE_, SIMD_WIDTH_AMD); -#endif -#ifdef CL_DEVICE_SIMD_INSTRUCTION_WIDTH_AMD - ADD_ATTR("device_info", DEVICE_, SIMD_INSTRUCTION_WIDTH_AMD); -#endif -#ifdef CL_DEVICE_WAVEFRONT_WIDTH_AMD - ADD_ATTR("device_info", DEVICE_, WAVEFRONT_WIDTH_AMD); -#endif -#ifdef CL_DEVICE_GLOBAL_MEM_CHANNELS_AMD - ADD_ATTR("device_info", DEVICE_, GLOBAL_MEM_CHANNELS_AMD); -#endif -#ifdef CL_DEVICE_GLOBAL_MEM_CHANNEL_BANKS_AMD - ADD_ATTR("device_info", DEVICE_, GLOBAL_MEM_CHANNEL_BANKS_AMD); -#endif -#ifdef CL_DEVICE_GLOBAL_MEM_CHANNEL_BANK_WIDTH_AMD - ADD_ATTR("device_info", DEVICE_, GLOBAL_MEM_CHANNEL_BANK_WIDTH_AMD); -#endif -#ifdef CL_DEVICE_LOCAL_MEM_SIZE_PER_COMPUTE_UNIT_AMD - ADD_ATTR("device_info", DEVICE_, LOCAL_MEM_SIZE_PER_COMPUTE_UNIT_AMD); -#endif -#ifdef CL_DEVICE_LOCAL_MEM_BANKS_AMD - ADD_ATTR("device_info", DEVICE_, LOCAL_MEM_BANKS_AMD); -#endif - -#ifdef CL_DEVICE_THREAD_TRACE_SUPPORTED_AMD - ADD_ATTR("device_info", DEVICE_, THREAD_TRACE_SUPPORTED_AMD); -#endif -#ifdef CL_DEVICE_GFXIP_MAJOR_AMD - ADD_ATTR("device_info", DEVICE_, GFXIP_MAJOR_AMD); -#endif -#ifdef CL_DEVICE_GFXIP_MINOR_AMD - ADD_ATTR("device_info", DEVICE_, GFXIP_MINOR_AMD); -#endif -#ifdef CL_DEVICE_AVAILABLE_ASYNC_QUEUES_AMD - ADD_ATTR("device_info", DEVICE_, AVAILABLE_ASYNC_QUEUES_AMD); -#endif - -#ifdef CL_DEVICE_MAX_ATOMIC_COUNTERS_EXT - ADD_ATTR("device_info", DEVICE_, MAX_ATOMIC_COUNTERS_EXT); -#endif -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("device_info", DEVICE_, LINKER_AVAILABLE); - ADD_ATTR("device_info", DEVICE_, BUILT_IN_KERNELS); - ADD_ATTR("device_info", DEVICE_, IMAGE_MAX_BUFFER_SIZE); - ADD_ATTR("device_info", DEVICE_, IMAGE_MAX_ARRAY_SIZE); - ADD_ATTR("device_info", DEVICE_, PARENT_DEVICE); - ADD_ATTR("device_info", DEVICE_, PARTITION_MAX_SUB_DEVICES); - ADD_ATTR("device_info", DEVICE_, PARTITION_PROPERTIES); - ADD_ATTR("device_info", DEVICE_, PARTITION_AFFINITY_DOMAIN); - ADD_ATTR("device_info", DEVICE_, PARTITION_TYPE); - ADD_ATTR("device_info", DEVICE_, REFERENCE_COUNT); - ADD_ATTR("device_info", DEVICE_, PREFERRED_INTEROP_USER_SYNC); - ADD_ATTR("device_info", DEVICE_, PRINTF_BUFFER_SIZE); -#endif -#ifdef cl_khr_image2d_from_buffer - ADD_ATTR("device_info", DEVICE_, IMAGE_PITCH_ALIGNMENT); - ADD_ATTR("device_info", DEVICE_, IMAGE_BASE_ADDRESS_ALIGNMENT); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("device_info", DEVICE_, MAX_READ_WRITE_IMAGE_ARGS); - ADD_ATTR("device_info", DEVICE_, MAX_GLOBAL_VARIABLE_SIZE); - ADD_ATTR("device_info", DEVICE_, QUEUE_ON_DEVICE_PROPERTIES); - ADD_ATTR("device_info", DEVICE_, QUEUE_ON_DEVICE_PREFERRED_SIZE); - ADD_ATTR("device_info", DEVICE_, QUEUE_ON_DEVICE_MAX_SIZE); - ADD_ATTR("device_info", DEVICE_, MAX_ON_DEVICE_QUEUES); - ADD_ATTR("device_info", DEVICE_, MAX_ON_DEVICE_EVENTS); - ADD_ATTR("device_info", DEVICE_, SVM_CAPABILITIES); - ADD_ATTR("device_info", DEVICE_, GLOBAL_VARIABLE_PREFERRED_TOTAL_SIZE); - ADD_ATTR("device_info", DEVICE_, MAX_PIPE_ARGS); - ADD_ATTR("device_info", DEVICE_, PIPE_MAX_ACTIVE_RESERVATIONS); - ADD_ATTR("device_info", DEVICE_, PIPE_MAX_PACKET_SIZE); - ADD_ATTR("device_info", DEVICE_, PREFERRED_PLATFORM_ATOMIC_ALIGNMENT); - ADD_ATTR("device_info", DEVICE_, PREFERRED_GLOBAL_ATOMIC_ALIGNMENT); - ADD_ATTR("device_info", DEVICE_, PREFERRED_LOCAL_ATOMIC_ALIGNMENT); -#endif -#if PYOPENCL_CL_VERSION >= 0x2010 - ADD_ATTR("device_info", DEVICE_, IL_VERSION); - ADD_ATTR("device_info", DEVICE_, MAX_NUM_SUB_GROUPS); - ADD_ATTR("device_info", DEVICE_, SUB_GROUP_INDEPENDENT_FORWARD_PROGRESS); -#endif - /* cl_intel_advanced_motion_estimation */ -#ifdef CL_DEVICE_ME_VERSION_INTEL - ADD_ATTR("device_info", DEVICE_, ME_VERSION_INTEL); -#endif - - /* cl_qcom_ext_host_ptr */ -#ifdef CL_DEVICE_EXT_MEM_PADDING_IN_BYTES_QCOM - ADD_ATTR("device_info", DEVICE_, EXT_MEM_PADDING_IN_BYTES_QCOM); -#endif -#ifdef CL_DEVICE_PAGE_SIZE_QCOM - ADD_ATTR("device_info", DEVICE_, PAGE_SIZE_QCOM); -#endif - - /* cl_khr_spir */ -#ifdef CL_DEVICE_SPIR_VERSIONS - ADD_ATTR("device_info", DEVICE_, SPIR_VERSIONS); -#endif - - /* cl_altera_device_temperature */ -#ifdef CL_DEVICE_CORE_TEMPERATURE_ALTERA - ADD_ATTR("device_info", DEVICE_, CORE_TEMPERATURE_ALTERA); -#endif - - /* cl_intel_simultaneous_sharing */ -#ifdef CL_DEVICE_SIMULTANEOUS_INTEROPS_INTEL - ADD_ATTR("device_info", DEVICE_, SIMULTANEOUS_INTEROPS_INTEL); -#endif -#ifdef CL_DEVICE_NUM_SIMULTANEOUS_INTEROPS_INTEL - ADD_ATTR("device_info", DEVICE_, NUM_SIMULTANEOUS_INTEROPS_INTEL); -#endif - - // device_fp_config - ADD_ATTR("device_fp_config", FP_, DENORM); - ADD_ATTR("device_fp_config", FP_, INF_NAN); - ADD_ATTR("device_fp_config", FP_, ROUND_TO_NEAREST); - ADD_ATTR("device_fp_config", FP_, ROUND_TO_ZERO); - ADD_ATTR("device_fp_config", FP_, ROUND_TO_INF); - ADD_ATTR("device_fp_config", FP_, FMA); -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("device_fp_config", FP_, SOFT_FLOAT); -#endif -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("device_fp_config", FP_, CORRECTLY_ROUNDED_DIVIDE_SQRT); -#endif - - - // device_mem_cache_type - ADD_ATTR("device_mem_cache_type", , NONE); - ADD_ATTR("device_mem_cache_type", , READ_ONLY_CACHE); - ADD_ATTR("device_mem_cache_type", , READ_WRITE_CACHE); - - - // device_local_mem_type - ADD_ATTR("device_local_mem_type", , LOCAL); - ADD_ATTR("device_local_mem_type", , GLOBAL); - - - // device_exec_capabilities - ADD_ATTR("device_exec_capabilities", EXEC_, KERNEL); - ADD_ATTR("device_exec_capabilities", EXEC_, NATIVE_KERNEL); -#ifdef CL_EXEC_IMMEDIATE_EXECUTION_INTEL - ADD_ATTR("device_exec_capabilities", EXEC_, IMMEDIATE_EXECUTION_INTEL); -#endif - -#if PYOPENCL_CL_VERSION >= 0x2000 - // device_svm_capabilities - ADD_ATTR("device_svm_capabilities", DEVICE_SVM_, COARSE_GRAIN_BUFFER); - ADD_ATTR("device_svm_capabilities", DEVICE_SVM_, FINE_GRAIN_BUFFER); - ADD_ATTR("device_svm_capabilities", DEVICE_SVM_, FINE_GRAIN_SYSTEM); - ADD_ATTR("device_svm_capabilities", DEVICE_SVM_, ATOMICS); -#endif - - - // command_queue_properties - ADD_ATTR("command_queue_properties", QUEUE_, OUT_OF_ORDER_EXEC_MODE_ENABLE); - ADD_ATTR("command_queue_properties", QUEUE_, PROFILING_ENABLE); -#ifdef CL_QUEUE_IMMEDIATE_EXECUTION_ENABLE_INTEL - ADD_ATTR("command_queue_properties", QUEUE_, IMMEDIATE_EXECUTION_ENABLE_INTEL); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("command_queue_properties", QUEUE_, ON_DEVICE); - ADD_ATTR("command_queue_properties", QUEUE_, ON_DEVICE_DEFAULT); -#endif - - - // context_info - ADD_ATTR("context_info", CONTEXT_, REFERENCE_COUNT); - ADD_ATTR("context_info", CONTEXT_, DEVICES); - ADD_ATTR("context_info", CONTEXT_, PROPERTIES); -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("context_info", CONTEXT_, NUM_DEVICES); -#endif -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("context_info", CONTEXT_, INTEROP_USER_SYNC); -#endif - - - // gl_context_info -#if defined(cl_khr_gl_sharing) && (cl_khr_gl_sharing >= 1) - ADD_ATTR("gl_context_info", , CURRENT_DEVICE_FOR_GL_CONTEXT_KHR); - ADD_ATTR("gl_context_info", , DEVICES_FOR_GL_CONTEXT_KHR); -#endif - - - // context_properties - ADD_ATTR("context_properties", CONTEXT_, PLATFORM); -#if defined(cl_khr_gl_sharing) && (cl_khr_gl_sharing >= 1) - ADD_ATTR("context_properties", ,GL_CONTEXT_KHR); - ADD_ATTR("context_properties", ,EGL_DISPLAY_KHR); - ADD_ATTR("context_properties", ,GLX_DISPLAY_KHR); - ADD_ATTR("context_properties", ,WGL_HDC_KHR); - ADD_ATTR("context_properties", ,CGL_SHAREGROUP_KHR); -#endif -#if defined(__APPLE__) && defined(HAVE_GL) && !defined(PYOPENCL_APPLE_USE_CL_H) - ADD_ATTR("context_properties", ,CONTEXT_PROPERTY_USE_CGL_SHAREGROUP_APPLE); -#endif /* __APPLE__ */ -#ifdef CL_CONTEXT_OFFLINE_DEVICES_AMD - ADD_ATTR("context_properties", CONTEXT_, OFFLINE_DEVICES_AMD); -#endif - - - // command_queue_info - ADD_ATTR("command_queue_info", QUEUE_, CONTEXT); - ADD_ATTR("command_queue_info", QUEUE_, DEVICE); - ADD_ATTR("command_queue_info", QUEUE_, REFERENCE_COUNT); - ADD_ATTR("command_queue_info", QUEUE_, PROPERTIES); - - - // queue_properties -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("queue_properties", QUEUE_, PROPERTIES); - ADD_ATTR("queue_properties", QUEUE_, SIZE); -#endif - - - // mem_flags - ADD_ATTR("mem_flags", MEM_, READ_WRITE); - ADD_ATTR("mem_flags", MEM_, WRITE_ONLY); - ADD_ATTR("mem_flags", MEM_, READ_ONLY); - ADD_ATTR("mem_flags", MEM_, USE_HOST_PTR); - ADD_ATTR("mem_flags", MEM_, ALLOC_HOST_PTR); - ADD_ATTR("mem_flags", MEM_, COPY_HOST_PTR); -#ifdef cl_amd_device_memory_flags - ADD_ATTR("mem_flags", MEM_, USE_PERSISTENT_MEM_AMD); -#endif -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("mem_flags", MEM_, HOST_WRITE_ONLY); - ADD_ATTR("mem_flags", MEM_, HOST_READ_ONLY); - ADD_ATTR("mem_flags", MEM_, HOST_NO_ACCESS); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("mem_flags", MEM_, KERNEL_READ_AND_WRITE); -#endif - -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("svm_mem_flags", MEM_, READ_WRITE); - ADD_ATTR("svm_mem_flags", MEM_, WRITE_ONLY); - ADD_ATTR("svm_mem_flags", MEM_, READ_ONLY); - ADD_ATTR("svm_mem_flags", MEM_, SVM_FINE_GRAIN_BUFFER); - ADD_ATTR("svm_mem_flags", MEM_, SVM_ATOMICS); -#endif - - - // channel_order - ADD_ATTR("channel_order", , R); - ADD_ATTR("channel_order", , A); - ADD_ATTR("channel_order", , RG); - ADD_ATTR("channel_order", , RA); - ADD_ATTR("channel_order", , RGB); - ADD_ATTR("channel_order", , RGBA); - ADD_ATTR("channel_order", , BGRA); - ADD_ATTR("channel_order", , INTENSITY); - ADD_ATTR("channel_order", , LUMINANCE); -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("channel_order", , Rx); - ADD_ATTR("channel_order", , RGx); - ADD_ATTR("channel_order", , RGBx); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("channel_order", , sRGB); - ADD_ATTR("channel_order", , sRGBx); - ADD_ATTR("channel_order", , sRGBA); - ADD_ATTR("channel_order", , sBGRA); - ADD_ATTR("channel_order", , ABGR); -#endif - - - // channel_type - ADD_ATTR("channel_type", , SNORM_INT8); - ADD_ATTR("channel_type", , SNORM_INT16); - ADD_ATTR("channel_type", , UNORM_INT8); - ADD_ATTR("channel_type", , UNORM_INT16); - ADD_ATTR("channel_type", , UNORM_SHORT_565); - ADD_ATTR("channel_type", , UNORM_SHORT_555); - ADD_ATTR("channel_type", , UNORM_INT_101010); - ADD_ATTR("channel_type", , SIGNED_INT8); - ADD_ATTR("channel_type", , SIGNED_INT16); - ADD_ATTR("channel_type", , SIGNED_INT32); - ADD_ATTR("channel_type", , UNSIGNED_INT8); - ADD_ATTR("channel_type", , UNSIGNED_INT16); - ADD_ATTR("channel_type", , UNSIGNED_INT32); - ADD_ATTR("channel_type", , HALF_FLOAT); - ADD_ATTR("channel_type", , FLOAT); - - - // mem_object_type - ADD_ATTR("mem_object_type", MEM_OBJECT_, BUFFER); - ADD_ATTR("mem_object_type", MEM_OBJECT_, IMAGE2D); - ADD_ATTR("mem_object_type", MEM_OBJECT_, IMAGE3D); -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("mem_object_type", MEM_OBJECT_, IMAGE2D_ARRAY); - ADD_ATTR("mem_object_type", MEM_OBJECT_, IMAGE1D); - ADD_ATTR("mem_object_type", MEM_OBJECT_, IMAGE1D_ARRAY); - ADD_ATTR("mem_object_type", MEM_OBJECT_, IMAGE1D_BUFFER); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("mem_object_type", MEM_OBJECT_, PIPE); -#endif - - - // mem_info - ADD_ATTR("mem_info", MEM_, TYPE); - ADD_ATTR("mem_info", MEM_, FLAGS); - ADD_ATTR("mem_info", MEM_, SIZE); - ADD_ATTR("mem_info", MEM_, HOST_PTR); - ADD_ATTR("mem_info", MEM_, MAP_COUNT); - ADD_ATTR("mem_info", MEM_, REFERENCE_COUNT); - ADD_ATTR("mem_info", MEM_, CONTEXT); -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("mem_info", MEM_, ASSOCIATED_MEMOBJECT); - ADD_ATTR("mem_info", MEM_, OFFSET); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("mem_info", MEM_, USES_SVM_POINTER); -#endif - - - // image_info - ADD_ATTR("image_info", IMAGE_, FORMAT); - ADD_ATTR("image_info", IMAGE_, ELEMENT_SIZE); - ADD_ATTR("image_info", IMAGE_, ROW_PITCH); - ADD_ATTR("image_info", IMAGE_, SLICE_PITCH); - ADD_ATTR("image_info", IMAGE_, WIDTH); - ADD_ATTR("image_info", IMAGE_, HEIGHT); - ADD_ATTR("image_info", IMAGE_, DEPTH); -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("image_info", IMAGE_, ARRAY_SIZE); - ADD_ATTR("image_info", IMAGE_, BUFFER); - ADD_ATTR("image_info", IMAGE_, NUM_MIP_LEVELS); - ADD_ATTR("image_info", IMAGE_, NUM_SAMPLES); -#endif - - - // addressing_mode - ADD_ATTR("addressing_mode", ADDRESS_, NONE); - ADD_ATTR("addressing_mode", ADDRESS_, CLAMP_TO_EDGE); - ADD_ATTR("addressing_mode", ADDRESS_, CLAMP); - ADD_ATTR("addressing_mode", ADDRESS_, REPEAT); -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("addressing_mode", ADDRESS_, MIRRORED_REPEAT); -#endif - - - // filter_mode - ADD_ATTR("filter_mode", FILTER_, NEAREST); - ADD_ATTR("filter_mode", FILTER_, LINEAR); - - - // sampler_info - ADD_ATTR("sampler_info", SAMPLER_, REFERENCE_COUNT); - ADD_ATTR("sampler_info", SAMPLER_, CONTEXT); - ADD_ATTR("sampler_info", SAMPLER_, NORMALIZED_COORDS); - ADD_ATTR("sampler_info", SAMPLER_, ADDRESSING_MODE); - ADD_ATTR("sampler_info", SAMPLER_, FILTER_MODE); -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("sampler_info", SAMPLER_, MIP_FILTER_MODE); - ADD_ATTR("sampler_info", SAMPLER_, LOD_MIN); - ADD_ATTR("sampler_info", SAMPLER_, LOD_MAX); -#endif - - - // map_flags - ADD_ATTR("map_flags", MAP_, READ); - ADD_ATTR("map_flags", MAP_, WRITE); -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("map_flags", MAP_, WRITE_INVALIDATE_REGION); -#endif - - - // program_info - ADD_ATTR("program_info", PROGRAM_, REFERENCE_COUNT); - ADD_ATTR("program_info", PROGRAM_, CONTEXT); - ADD_ATTR("program_info", PROGRAM_, NUM_DEVICES); - ADD_ATTR("program_info", PROGRAM_, DEVICES); - ADD_ATTR("program_info", PROGRAM_, SOURCE); - ADD_ATTR("program_info", PROGRAM_, BINARY_SIZES); - ADD_ATTR("program_info", PROGRAM_, BINARIES); -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("program_info", PROGRAM_, NUM_KERNELS); - ADD_ATTR("program_info", PROGRAM_, KERNEL_NAMES); -#endif - - - // program_build_info - ADD_ATTR("program_build_info", PROGRAM_BUILD_, STATUS); - ADD_ATTR("program_build_info", PROGRAM_BUILD_, OPTIONS); - ADD_ATTR("program_build_info", PROGRAM_BUILD_, LOG); -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("program_build_info", PROGRAM_, BINARY_TYPE); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("program_build_info", PROGRAM_BUILD_, GLOBAL_VARIABLE_TOTAL_SIZE); -#endif - - - // program_binary_type -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("program_binary_type", PROGRAM_BINARY_TYPE_, NONE); - ADD_ATTR("program_binary_type", PROGRAM_BINARY_TYPE_, COMPILED_OBJECT); - ADD_ATTR("program_binary_type", PROGRAM_BINARY_TYPE_, LIBRARY); - ADD_ATTR("program_binary_type", PROGRAM_BINARY_TYPE_, EXECUTABLE); -#endif - - - // kernel_info - ADD_ATTR("kernel_info", KERNEL_, FUNCTION_NAME); - ADD_ATTR("kernel_info", KERNEL_, NUM_ARGS); - ADD_ATTR("kernel_info", KERNEL_, REFERENCE_COUNT); - ADD_ATTR("kernel_info", KERNEL_, CONTEXT); - ADD_ATTR("kernel_info", KERNEL_, PROGRAM); -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("kernel_info", KERNEL_, ATTRIBUTES); -#endif - - - // kernel_arg_info -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("kernel_arg_info", KERNEL_ARG_, ADDRESS_QUALIFIER); - ADD_ATTR("kernel_arg_info", KERNEL_ARG_, ACCESS_QUALIFIER); - ADD_ATTR("kernel_arg_info", KERNEL_ARG_, TYPE_NAME); - ADD_ATTR("kernel_arg_info", KERNEL_ARG_, TYPE_QUALIFIER); - ADD_ATTR("kernel_arg_info", KERNEL_ARG_, NAME); -#endif - - - // kernel_arg_address_qualifier -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("kernel_arg_address_qualifier", KERNEL_ARG_ADDRESS_, GLOBAL); - ADD_ATTR("kernel_arg_address_qualifier", KERNEL_ARG_ADDRESS_, LOCAL); - ADD_ATTR("kernel_arg_address_qualifier", KERNEL_ARG_ADDRESS_, CONSTANT); - ADD_ATTR("kernel_arg_address_qualifier", KERNEL_ARG_ADDRESS_, PRIVATE); -#endif - - - // kernel_arg_access_qualifier -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("kernel_arg_access_qualifier", KERNEL_ARG_ACCESS_, READ_ONLY); - ADD_ATTR("kernel_arg_access_qualifier", KERNEL_ARG_ACCESS_, WRITE_ONLY); - ADD_ATTR("kernel_arg_access_qualifier", KERNEL_ARG_ACCESS_, READ_WRITE); - ADD_ATTR("kernel_arg_access_qualifier", KERNEL_ARG_ACCESS_, NONE); -#endif - - - // kernel_arg_type_qualifier -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("kernel_arg_type_qualifier", KERNEL_ARG_TYPE_, NONE); - ADD_ATTR("kernel_arg_type_qualifier", KERNEL_ARG_TYPE_, CONST); - ADD_ATTR("kernel_arg_type_qualifier", KERNEL_ARG_TYPE_, RESTRICT); - ADD_ATTR("kernel_arg_type_qualifier", KERNEL_ARG_TYPE_, VOLATILE); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("kernel_arg_type_qualifier", KERNEL_ARG_TYPE_, PIPE); -#endif - - - // kernel_work_group_info - ADD_ATTR("kernel_work_group_info", KERNEL_, WORK_GROUP_SIZE); - ADD_ATTR("kernel_work_group_info", KERNEL_, COMPILE_WORK_GROUP_SIZE); - ADD_ATTR("kernel_work_group_info", KERNEL_, LOCAL_MEM_SIZE); -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("kernel_work_group_info", KERNEL_, PREFERRED_WORK_GROUP_SIZE_MULTIPLE); - ADD_ATTR("kernel_work_group_info", KERNEL_, PRIVATE_MEM_SIZE); -#endif -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("kernel_work_group_info", KERNEL_, GLOBAL_WORK_SIZE); -#endif - - - // event_info - ADD_ATTR("event_info", EVENT_, COMMAND_QUEUE); - ADD_ATTR("event_info", EVENT_, COMMAND_TYPE); - ADD_ATTR("event_info", EVENT_, REFERENCE_COUNT); - ADD_ATTR("event_info", EVENT_, COMMAND_EXECUTION_STATUS); -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("event_info", EVENT_, CONTEXT); -#endif - - - // command_type - ADD_ATTR("command_type", COMMAND_, NDRANGE_KERNEL); - ADD_ATTR("command_type", COMMAND_, TASK); - ADD_ATTR("command_type", COMMAND_, NATIVE_KERNEL); - ADD_ATTR("command_type", COMMAND_, READ_BUFFER); - ADD_ATTR("command_type", COMMAND_, WRITE_BUFFER); - ADD_ATTR("command_type", COMMAND_, COPY_BUFFER); - ADD_ATTR("command_type", COMMAND_, READ_IMAGE); - ADD_ATTR("command_type", COMMAND_, WRITE_IMAGE); - ADD_ATTR("command_type", COMMAND_, COPY_IMAGE); - ADD_ATTR("command_type", COMMAND_, COPY_IMAGE_TO_BUFFER); - ADD_ATTR("command_type", COMMAND_, COPY_BUFFER_TO_IMAGE); - ADD_ATTR("command_type", COMMAND_, MAP_BUFFER); - ADD_ATTR("command_type", COMMAND_, MAP_IMAGE); - ADD_ATTR("command_type", COMMAND_, UNMAP_MEM_OBJECT); - ADD_ATTR("command_type", COMMAND_, MARKER); - ADD_ATTR("command_type", COMMAND_, ACQUIRE_GL_OBJECTS); - ADD_ATTR("command_type", COMMAND_, RELEASE_GL_OBJECTS); -#if PYOPENCL_CL_VERSION >= 0x1010 - ADD_ATTR("command_type", COMMAND_, READ_BUFFER_RECT); - ADD_ATTR("command_type", COMMAND_, WRITE_BUFFER_RECT); - ADD_ATTR("command_type", COMMAND_, COPY_BUFFER_RECT); - ADD_ATTR("command_type", COMMAND_, USER); -#endif -#ifdef cl_ext_migrate_memobject - ADD_ATTR("command_type", COMMAND_, MIGRATE_MEM_OBJECT_EXT); -#endif -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("command_type", COMMAND_, BARRIER); - ADD_ATTR("command_type", COMMAND_, MIGRATE_MEM_OBJECTS); - ADD_ATTR("command_type", COMMAND_, FILL_BUFFER); - ADD_ATTR("command_type", COMMAND_, FILL_IMAGE); -#endif -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("command_type", COMMAND_, SVM_FREE); - ADD_ATTR("command_type", COMMAND_, SVM_MEMCPY); - ADD_ATTR("command_type", COMMAND_, SVM_MEMFILL); - ADD_ATTR("command_type", COMMAND_, SVM_MAP); - ADD_ATTR("command_type", COMMAND_, SVM_UNMAP); -#endif - - - // command_execution_status - ADD_ATTR("command_execution_status", , COMPLETE); - ADD_ATTR("command_execution_status", , RUNNING); - ADD_ATTR("command_execution_status", , SUBMITTED); - ADD_ATTR("command_execution_status", , QUEUED); - - - // profiling_info - ADD_ATTR("profiling_info", PROFILING_COMMAND_, QUEUED); - ADD_ATTR("profiling_info", PROFILING_COMMAND_, SUBMIT); - ADD_ATTR("profiling_info", PROFILING_COMMAND_, START); - ADD_ATTR("profiling_info", PROFILING_COMMAND_, END); -#if PYOPENCL_CL_VERSION >= 0x2000 - ADD_ATTR("profiling_info", PROFILING_COMMAND_, COMPLETE); -#endif - - - // mem_migration_flags -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("mem_migration_flags", MIGRATE_MEM_OBJECT_, HOST); - ADD_ATTR("mem_migration_flags", MIGRATE_MEM_OBJECT_, CONTENT_UNDEFINED); -#endif - - - // mem_migration_flags_ext -#ifdef cl_ext_migrate_memobject - ADD_ATTR("mem_migration_flags_ext", MIGRATE_MEM_OBJECT_, HOST, _EXT); - - // As of 2018-07-11, the official headers seem to have dropped this: -#ifdef CL_MIGRATE_MEM_OBJECT_CONTENT_UNDEFINED_EXT - ADD_ATTR("mem_migration_flags_ext", MIGRATE_MEM_OBJECT_, - CONTENT_UNDEFINED, _EXT); -#endif - -#endif - - - // device_partition_property -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("device_partition_property", DEVICE_PARTITION_, EQUALLY); - ADD_ATTR("device_partition_property", DEVICE_PARTITION_, BY_COUNTS); - ADD_ATTR("device_partition_property", DEVICE_PARTITION_, BY_COUNTS_LIST_END); - ADD_ATTR("device_partition_property", DEVICE_PARTITION_, BY_AFFINITY_DOMAIN); -#endif - - - // device_affinity_domain -#if PYOPENCL_CL_VERSION >= 0x1020 - ADD_ATTR("device_affinity_domain", DEVICE_AFFINITY_DOMAIN_, NUMA); - ADD_ATTR("device_affinity_domain", DEVICE_AFFINITY_DOMAIN_, L4_CACHE); - ADD_ATTR("device_affinity_domain", DEVICE_AFFINITY_DOMAIN_, L3_CACHE); - ADD_ATTR("device_affinity_domain", DEVICE_AFFINITY_DOMAIN_, L2_CACHE); - ADD_ATTR("device_affinity_domain", DEVICE_AFFINITY_DOMAIN_, L1_CACHE); - ADD_ATTR("device_affinity_domain", DEVICE_AFFINITY_DOMAIN_, - NEXT_PARTITIONABLE); -#endif - - -#ifdef HAVE_GL - // gl_object_type - ADD_ATTR("gl_object_type", GL_OBJECT_, BUFFER); - ADD_ATTR("gl_object_type", GL_OBJECT_, TEXTURE2D); - ADD_ATTR("gl_object_type", GL_OBJECT_, TEXTURE3D); - ADD_ATTR("gl_object_type", GL_OBJECT_, RENDERBUFFER); - - - // gl_texture_info - ADD_ATTR("gl_texture_info", GL_, TEXTURE_TARGET); - ADD_ATTR("gl_texture_info", GL_, MIPMAP_LEVEL); -#endif - - - // migrate_mem_object_flags_ext -#ifdef cl_ext_migrate_memobject - ADD_ATTR("migrate_mem_object_flags_ext", MIGRATE_MEM_OBJECT_, HOST, _EXT); -#endif -} -- GitLab