--- OpenCL/OpenCL.pm	2012/05/03 23:32:47	1.70
+++ OpenCL/OpenCL.pm	2012/05/05 14:17:39	1.81
@@ -30,8 +30,8 @@
 which represents basically a function call with argument values.
 
 OpenCL::Memory objects of various flavours: OpenCL::Buffer objects (flat
-memory areas, think arrays or structs) and OpenCL::Image objects (think 2d
-or 3d array) for bulk data and input and output for kernels.
+memory areas, think arrays or structs) and OpenCL::Image objects (think 2D
+or 3D array) for bulk data and input and output for kernels.
 
 OpenCL::Sampler objects, which are kind of like texture filter modes in
 OpenGL.
@@ -45,7 +45,8 @@
 
 =head2 HELPFUL RESOURCES
 
-The OpenCL specs used to develop this module:
+The OpenCL specs used to develop this module - download these and keept
+hema round, they are required reference material:
 
    http://www.khronos.org/registry/cl/specs/opencl-1.1.pdf
    http://www.khronos.org/registry/cl/specs/opencl-1.2.pdf
@@ -212,10 +213,10 @@
    $ev->wait;
 
 =head2 Use the OpenGL module to share a texture between OpenCL and OpenGL and draw some julia
-set tunnel effect.
+set flight effect.
 
-This is quite a long example to get you going - you can download it from
-L<http://cvs.schmorp.de/OpenCL/examples/juliaflight>.
+This is quite a long example to get you going - you can also download it
+from L<http://cvs.schmorp.de/OpenCL/examples/juliaflight>.
 
    use OpenGL ":all";
    use OpenCL;
@@ -322,7 +323,7 @@
 
    my $tex = $ctx->image2d (OpenCL::MEM_WRITE_ONLY, OpenCL::RGBA, OpenCL::UNORM_INT8, $S, $S);
 
-And in the darw loop, intead of acquire_gl_objects/release_gl_objects, you
+And in the draw loop, intead of acquire_gl_objects/release_gl_objects, you
 would read the image2d after the kernel has written it:
 
    $queue->read_image ($tex, 0, 0, 0, 0, $S, $S, 1, 0, 0, my $data);
@@ -334,6 +335,23 @@
 The fully modified example can be found at
 L<http://cvs.schmorp.de/OpenCL/examples/juliaflight-nosharing>.
 
+=head2 Julia sets look soooo 80ies.
+
+Then colour them differently, e.g. using orbit traps! Replace the loop and
+colour calculation from the previous examples by this:
+
+  float2 dm = (float2)(1.f, 1.f);
+
+  for (int i = 0; i < 25; ++i)
+    {
+      z = (float2)(z.x * z.x - z.y * z.y, 2.f * z.x * z.y) + c;
+      dm = fmin (dm, (float2)(fabs (dot (z, z) - 1.f), fabs (z.x - 1.f)));
+    }
+
+  float3 colour = (float3)(dm.x * dm.y, dm.x * dm.y, dm.x);
+
+Also try C<-10.f> instead of C<-1.f>.
+
 =head1 DOCUMENTATION
 
 =head2 BASIC CONVENTIONS
@@ -372,12 +390,53 @@
 context, no event is created. In all other contexts an event is returned
 by the method.
 
-=item * This module expects all functions to return C<CL_SUCCESS>. If any
+=item * This module expects all functions to return C<OpenCL::SUCCESS>. If any
 other status is returned the function will throw an exception, so you
 don't normally have to to any error checking.
 
 =back
 
+=head2 CONSTANTS
+
+All C<CL_xxx> constants that this module supports are always available
+in the C<OpenCL> namespace as C<OpenCL::xxx> (i.e. without the C<CL_>
+prefix). Constants which are not defined in the header files used during
+compilation, or otherwise are not available, will have the value C<-1>.
+
+The latest version of this module knows and exports the constants
+listed in L<http://cvs.schmorp.de/OpenCL/constiv.h>.
+
+=head2 OPENCL 1.1 VS. OPENCL 1.2
+
+This module supports both OpenCL version 1.1 and 1.2, although the OpenCL
+1.2 interface hasn't been tested much for lack of availability of an
+actual implementation.
+
+Every function or method in this manual page that interfaces to a
+particular OpenCL function has a link to the its C manual page.
+
+If the link contains a F<1.1>, then this function is an OpenCL 1.1
+function. Most but not all also exist in OpenCL 1.2, and this module
+tries to emulate the missing ones for you, when told to do so at
+compiletime. You can check whether a function was removed in OpenCL 1.2 by
+replacing the F<1.1> component in the URL by F<1.2>.
+
+If the link contains a F<1.2>, then this is a OpenCL 1.2-only
+function. Even if the module was compiled with OpenCL 1.2 header files
+and has an 1.2 OpenCL library, calling such a function on a platform that
+doesn't implement 1.2 causes undefined behaviour, usually a crash (But
+this is not guaranteed).
+
+You can find out whether this module was compiled to prefer 1.1
+functionality by ooking at C<OpenCL::PREFER_1_1> - if it is true, then
+1.1 functions generally are implemented using 1.1 OpenCL functions. If it
+is false, then 1.1 functions missing from 1.2 are emulated by calling 1.2
+fucntions.
+
+This is a somewhat sorry state of affairs, but the Khronos group choose to
+make every release of OpenCL source and binary incompatible with previous
+releases.
+
 =head2 PERL AND OPENCL TYPES
 
 This handy(?) table lists OpenCL types and their perl, PDL and pack/unpack
@@ -418,13 +477,20 @@
 perhaps most important, to signal completion of queued jobs (by setting
 callbacks on OpenCL::Event objects).
 
+The OpenCL module converts all these callbacks into events - you can
+still register callbacks, but they are not executed when your OpenCL
+implementation calls the actual callback, but only later. Therefore, none
+of the limitations of OpenCL callbacks apply to the perl implementation:
+it is perfectly safe to make blocking operations from event callbacks, and
+enqueued operations don't need to be flushed.
+
 To facilitate this, this module maintains an event queue - each
 time an asynchronous event happens, it is queued, and perl will be
 interrupted. This is implemented via the L<Async::Interrupt> module. In
 addition, this module has L<AnyEvent> support, so it can seamlessly
 integrate itself into many event loops.
 
-Since this module is a bit hard to understand, here are some case examples:
+Since L<Async::Interrupt> is a bit hard to understand, here are some case examples:
 
 =head3 Don't use callbacks.
 
@@ -546,6 +612,8 @@
    @OpenCL::Kernel::ISA        =
    @OpenCL::Event::ISA         = OpenCL::Object::;
 
+   @OpenCL::SubDevice::ISA     = OpenCL::Device::;
+
    @OpenCL::Buffer::ISA        =
    @OpenCL::Image::ISA         = OpenCL::Memory::;
 
@@ -575,9 +643,12 @@
 
 =item $str = OpenCL::err2str [$errval]
 
-Converts an error value into a human readable string. IF no error value is
+Converts an error value into a human readable string. If no error value is
 given, then the last error will be used (as returned by OpenCL::errno).
 
+The latest version of this module knows the error constants
+listed in L<http://cvs.schmorp.de/OpenCL/errstr.h>.
+
 =item $str = OpenCL::enum2str $enum
 
 Converts most enum values (of parameter names, image format constants,
@@ -586,6 +657,9 @@
 very helpful to pass it through this function to maybe get some readable
 string out of it.
 
+The latest version of this module knows the enumaration constants
+listed in L<http://cvs.schmorp.de/OpenCL/errstr.h>.
+
 =item @platforms = OpenCL::platforms
 
 Returns all available OpenCL::Platform objects.
@@ -595,13 +669,17 @@
 =item $ctx = OpenCL::context_from_type $properties, $type = OpenCL::DEVICE_TYPE_DEFAULT, $callback->($err, $pvt) = $print_stderr
 
 Tries to create a context from a default device and platform type - never worked for me.
+Consider using C<< $platform->context_from_type >> instead.
+
+type: OpenCL::DEVICE_TYPE_DEFAULT, OpenCL::DEVICE_TYPE_CPU, OpenCL::DEVICE_TYPE_GPU,
+OpenCL::DEVICE_TYPE_ACCELERATOR, OpenCL::DEVICE_TYPE_CUSTOM, OpenCL::DEVICE_TYPE_ALL.
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContextFromType.html>
 
 =item $ctx = OpenCL::context $properties, \@devices, $callback->($err, $pvt) = $print_stderr)
 
-Create a new OpenCL::Context object using the given device object(s). This
-function isn't implemented yet, use C<< $platform->context >> instead.
+Create a new OpenCL::Context object using the given device object(s).
+Consider using C<< $platform->context >> instead.
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContext.html>
 
@@ -692,7 +770,7 @@
 =item $ctx = $platform->context ($properties, \@devices, $callback->($err, $pvt) = $print_stderr)
 
 Create a new OpenCL::Context object using the given device object(s)- a
-CL_CONTEXT_PLATFORM property is supplied automatically.
+OpenCL::CONTEXT_PLATFORM property is supplied automatically.
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContext.html>
 
@@ -718,23 +796,23 @@
 
 =item $string = $platform->profile
 
-Calls C<clGetPlatformInfo> with C<CL_PLATFORM_PROFILE> and returns the result.
+Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_PROFILE> and returns the result.
 
 =item $string = $platform->version
 
-Calls C<clGetPlatformInfo> with C<CL_PLATFORM_VERSION> and returns the result.
+Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_VERSION> and returns the result.
 
 =item $string = $platform->name
 
-Calls C<clGetPlatformInfo> with C<CL_PLATFORM_NAME> and returns the result.
+Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_NAME> and returns the result.
 
 =item $string = $platform->vendor
 
-Calls C<clGetPlatformInfo> with C<CL_PLATFORM_VENDOR> and returns the result.
+Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_VENDOR> and returns the result.
 
 =item $string = $platform->extensions
 
-Calls C<clGetPlatformInfo> with C<CL_PLATFORM_EXTENSIONS> and returns the result.
+Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_EXTENSIONS> and returns the result.
 
 =for gengetinfo end platform
 
@@ -748,273 +826,305 @@
 
 See C<< $platform->info >> for details.
 
+type: OpenCL::DEVICE_TYPE_DEFAULT, OpenCL::DEVICE_TYPE_CPU,
+OpenCL::DEVICE_TYPE_GPU, OpenCL::DEVICE_TYPE_ACCELERATOR,
+OpenCL::DEVICE_TYPE_CUSTOM, OpenCL::DEVICE_TYPE_ALL.
+
+fp_config: OpenCL::FP_DENORM, OpenCL::FP_INF_NAN, OpenCL::FP_ROUND_TO_NEAREST,
+OpenCL::FP_ROUND_TO_ZERO, OpenCL::FP_ROUND_TO_INF, OpenCL::FP_FMA,
+OpenCL::FP_SOFT_FLOAT, OpenCL::FP_CORRECTLY_ROUNDED_DIVIDE_SQRT.
+
+mem_cache_type: OpenCL::NONE, OpenCL::READ_ONLY_CACHE, OpenCL::READ_WRITE_CACHE.
+
+local_mem_type: OpenCL::LOCAL, OpenCL::GLOBAL.
+
+exec_capabilities: OpenCL::EXEC_KERNEL, OpenCL::EXEC_NATIVE_KERNEL.
+
+command_queue_properties: OpenCL::QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE,
+OpenCL::QUEUE_PROFILING_ENABLE.
+
+partition_properties: OpenCL::DEVICE_PARTITION_EQUALLY,
+OpenCL::DEVICE_PARTITION_BY_COUNTS, OpenCL::DEVICE_PARTITION_BY_COUNTS_LIST_END,
+OpenCL::DEVICE_PARTITION_BY_AFFINITY_DOMAIN.
+
+affinity_domain: OpenCL::DEVICE_AFFINITY_DOMAIN_NUMA,
+OpenCL::DEVICE_AFFINITY_DOMAIN_L4_CACHE, OpenCL::DEVICE_AFFINITY_DOMAIN_L3_CACHE,
+OpenCL::DEVICE_AFFINITY_DOMAIN_L2_CACHE, OpenCL::DEVICE_AFFINITY_DOMAIN_L1_CACHE,
+OpenCL::DEVICE_AFFINITY_DOMAIN_NEXT_PARTITIONABLE.
+
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetDeviceInfo.html>
 
+=item @devices = $device->sub_devices (\@properties)
+
+Creates OpencL::SubDevice objects by partitioning an existing device.
+
+L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateSubDevices.html>
+
 =for gengetinfo begin device
 
 =item $device_type = $device->type
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_TYPE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_TYPE> and returns the result.
 
 =item $uint = $device->vendor_id
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_VENDOR_ID> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_VENDOR_ID> and returns the result.
 
 =item $uint = $device->max_compute_units
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_COMPUTE_UNITS> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_COMPUTE_UNITS> and returns the result.
 
 =item $uint = $device->max_work_item_dimensions
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_WORK_ITEM_DIMENSIONS> and returns the result.
 
 =item $int = $device->max_work_group_size
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_WORK_GROUP_SIZE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_WORK_GROUP_SIZE> and returns the result.
 
 =item @ints = $device->max_work_item_sizes
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_WORK_ITEM_SIZES> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_WORK_ITEM_SIZES> and returns the result.
 
 =item $uint = $device->preferred_vector_width_char
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PREFERRED_VECTOR_WIDTH_CHAR> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_CHAR> and returns the result.
 
 =item $uint = $device->preferred_vector_width_short
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PREFERRED_VECTOR_WIDTH_SHORT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_SHORT> and returns the result.
 
 =item $uint = $device->preferred_vector_width_int
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PREFERRED_VECTOR_WIDTH_INT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_INT> and returns the result.
 
 =item $uint = $device->preferred_vector_width_long
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PREFERRED_VECTOR_WIDTH_LONG> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_LONG> and returns the result.
 
 =item $uint = $device->preferred_vector_width_float
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT> and returns the result.
 
 =item $uint = $device->preferred_vector_width_double
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE> and returns the result.
 
 =item $uint = $device->max_clock_frequency
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_CLOCK_FREQUENCY> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_CLOCK_FREQUENCY> and returns the result.
 
 =item $bitfield = $device->address_bits
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_ADDRESS_BITS> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_ADDRESS_BITS> and returns the result.
 
 =item $uint = $device->max_read_image_args
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_READ_IMAGE_ARGS> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_READ_IMAGE_ARGS> and returns the result.
 
 =item $uint = $device->max_write_image_args
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_WRITE_IMAGE_ARGS> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_WRITE_IMAGE_ARGS> and returns the result.
 
 =item $ulong = $device->max_mem_alloc_size
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_MEM_ALLOC_SIZE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_MEM_ALLOC_SIZE> and returns the result.
 
 =item $int = $device->image2d_max_width
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_IMAGE2D_MAX_WIDTH> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE2D_MAX_WIDTH> and returns the result.
 
 =item $int = $device->image2d_max_height
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_IMAGE2D_MAX_HEIGHT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE2D_MAX_HEIGHT> and returns the result.
 
 =item $int = $device->image3d_max_width
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_IMAGE3D_MAX_WIDTH> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE3D_MAX_WIDTH> and returns the result.
 
 =item $int = $device->image3d_max_height
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_IMAGE3D_MAX_HEIGHT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE3D_MAX_HEIGHT> and returns the result.
 
 =item $int = $device->image3d_max_depth
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_IMAGE3D_MAX_DEPTH> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE3D_MAX_DEPTH> and returns the result.
 
 =item $uint = $device->image_support
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_IMAGE_SUPPORT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE_SUPPORT> and returns the result.
 
 =item $int = $device->max_parameter_size
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_PARAMETER_SIZE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_PARAMETER_SIZE> and returns the result.
 
 =item $uint = $device->max_samplers
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_SAMPLERS> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_SAMPLERS> and returns the result.
 
 =item $uint = $device->mem_base_addr_align
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MEM_BASE_ADDR_ALIGN> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MEM_BASE_ADDR_ALIGN> and returns the result.
 
 =item $uint = $device->min_data_type_align_size
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MIN_DATA_TYPE_ALIGN_SIZE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MIN_DATA_TYPE_ALIGN_SIZE> and returns the result.
 
 =item $device_fp_config = $device->single_fp_config
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_SINGLE_FP_CONFIG> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_SINGLE_FP_CONFIG> and returns the result.
 
 =item $device_mem_cache_type = $device->global_mem_cache_type
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_GLOBAL_MEM_CACHE_TYPE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_GLOBAL_MEM_CACHE_TYPE> and returns the result.
 
 =item $uint = $device->global_mem_cacheline_size
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_GLOBAL_MEM_CACHELINE_SIZE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_GLOBAL_MEM_CACHELINE_SIZE> and returns the result.
 
 =item $ulong = $device->global_mem_cache_size
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_GLOBAL_MEM_CACHE_SIZE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_GLOBAL_MEM_CACHE_SIZE> and returns the result.
 
 =item $ulong = $device->global_mem_size
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_GLOBAL_MEM_SIZE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_GLOBAL_MEM_SIZE> and returns the result.
 
 =item $ulong = $device->max_constant_buffer_size
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_CONSTANT_BUFFER_SIZE> and returns the result.
 
 =item $uint = $device->max_constant_args
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_CONSTANT_ARGS> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_CONSTANT_ARGS> and returns the result.
 
 =item $device_local_mem_type = $device->local_mem_type
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_LOCAL_MEM_TYPE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_LOCAL_MEM_TYPE> and returns the result.
 
 =item $ulong = $device->local_mem_size
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_LOCAL_MEM_SIZE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_LOCAL_MEM_SIZE> and returns the result.
 
 =item $boolean = $device->error_correction_support
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_ERROR_CORRECTION_SUPPORT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_ERROR_CORRECTION_SUPPORT> and returns the result.
 
 =item $int = $device->profiling_timer_resolution
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PROFILING_TIMER_RESOLUTION> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PROFILING_TIMER_RESOLUTION> and returns the result.
 
 =item $boolean = $device->endian_little
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_ENDIAN_LITTLE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_ENDIAN_LITTLE> and returns the result.
 
 =item $boolean = $device->available
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_AVAILABLE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_AVAILABLE> and returns the result.
 
 =item $boolean = $device->compiler_available
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_COMPILER_AVAILABLE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_COMPILER_AVAILABLE> and returns the result.
 
 =item $device_exec_capabilities = $device->execution_capabilities
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_EXECUTION_CAPABILITIES> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_EXECUTION_CAPABILITIES> and returns the result.
 
 =item $command_queue_properties = $device->properties
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_QUEUE_PROPERTIES> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_QUEUE_PROPERTIES> and returns the result.
 
 =item $ = $device->platform
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PLATFORM> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PLATFORM> and returns the result.
 
 =item $string = $device->name
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_NAME> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NAME> and returns the result.
 
 =item $string = $device->vendor
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_VENDOR> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_VENDOR> and returns the result.
 
 =item $string = $device->driver_version
 
-Calls C<clGetDeviceInfo> with C<CL_DRIVER_VERSION> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DRIVER_VERSION> and returns the result.
 
 =item $string = $device->profile
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PROFILE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PROFILE> and returns the result.
 
 =item $string = $device->version
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_VERSION> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_VERSION> and returns the result.
 
 =item $string = $device->extensions
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_EXTENSIONS> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_EXTENSIONS> and returns the result.
 
 =item $uint = $device->preferred_vector_width_half
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PREFERRED_VECTOR_WIDTH_HALF> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_HALF> and returns the result.
 
 =item $uint = $device->native_vector_width_char
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_NATIVE_VECTOR_WIDTH_CHAR> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_CHAR> and returns the result.
 
 =item $uint = $device->native_vector_width_short
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_NATIVE_VECTOR_WIDTH_SHORT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_SHORT> and returns the result.
 
 =item $uint = $device->native_vector_width_int
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_NATIVE_VECTOR_WIDTH_INT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_INT> and returns the result.
 
 =item $uint = $device->native_vector_width_long
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_NATIVE_VECTOR_WIDTH_LONG> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_LONG> and returns the result.
 
 =item $uint = $device->native_vector_width_float
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_NATIVE_VECTOR_WIDTH_FLOAT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_FLOAT> and returns the result.
 
 =item $uint = $device->native_vector_width_double
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_NATIVE_VECTOR_WIDTH_DOUBLE> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_DOUBLE> and returns the result.
 
 =item $uint = $device->native_vector_width_half
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_NATIVE_VECTOR_WIDTH_HALF> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_HALF> and returns the result.
 
 =item $device_fp_config = $device->double_fp_config
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_DOUBLE_FP_CONFIG> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_DOUBLE_FP_CONFIG> and returns the result.
 
 =item $device_fp_config = $device->half_fp_config
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_HALF_FP_CONFIG> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_HALF_FP_CONFIG> and returns the result.
 
 =item $boolean = $device->host_unified_memory
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_HOST_UNIFIED_MEMORY> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_HOST_UNIFIED_MEMORY> and returns the result.
 
 =item $device = $device->parent_device_ext
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PARENT_DEVICE_EXT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PARENT_DEVICE_EXT> and returns the result.
 
 =item @device_partition_property_exts = $device->partition_types_ext
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PARTITION_TYPES_EXT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PARTITION_TYPES_EXT> and returns the result.
 
 =item @device_partition_property_exts = $device->affinity_domains_ext
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_AFFINITY_DOMAINS_EXT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_AFFINITY_DOMAINS_EXT> and returns the result.
 
 =item $uint = $device->reference_count_ext
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_REFERENCE_COUNT_EXT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_REFERENCE_COUNT_EXT> and returns the result.
 
 =item @device_partition_property_exts = $device->partition_style_ext
 
-Calls C<clGetDeviceInfo> with C<CL_DEVICE_PARTITION_STYLE_EXT> and returns the result.
+Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PARTITION_STYLE_EXT> and returns the result.
 
 =for gengetinfo end device
 
@@ -1022,6 +1132,32 @@
 
 =head2 THE OpenCL::Context CLASS
 
+An OpenCL::Context is basically a container, or manager, for a number of
+devices of a platform. It is used to create all sorts of secondary objects
+such as buffers, queues, programs and so on.
+
+All context creation functions and methods take a list of properties
+(type-value pairs). All property values can be specified as integers -
+some additionally support other types:
+
+=over 4
+
+=item OpenCL::CONTEXT_PLATFORM
+
+Also accepts OpenCL::Platform objects.
+
+=item OpenCL::GLX_DISPLAY_KHR
+
+Also accepts C<undef>, in which case a deep and troubling hack is engaged
+to find the current glx display (see L<GLX SUPPORT>).
+
+=item OpenCL::GL_CONTEXT_KHR
+
+Also accepts C<undef>, in which case a deep and troubling hack is engaged
+to find the current glx context (see L<GLX SUPPORT>).
+
+=back
+
 =over 4
 
 =item $prog = $ctx->build_program ($program, $options = "")
@@ -1078,6 +1214,10 @@
 Creates a new OpenCL::Buffer (actually OpenCL::BufferObj) object with the
 given flags and octet-size.
 
+flags: OpenCL::MEM_READ_WRITE, OpenCL::MEM_WRITE_ONLY, OpenCL::MEM_READ_ONLY,
+OpenCL::MEM_USE_HOST_PTR, OpenCL::MEM_ALLOC_HOST_PTR, OpenCL::MEM_COPY_HOST_PTR,
+OpenCL::MEM_HOST_WRITE_ONLY, OpenCL::MEM_HOST_READ_ONLY, OpenCL::MEM_HOST_NO_ACCESS.
+
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateBuffer.html>
 
 =item $buf = $ctx->buffer_sv ($flags, $data)
@@ -1090,6 +1230,21 @@
 Creates a new OpenCL::Image object and optionally initialises it with
 the given data values.
 
+channel_order: OpenCL::R, OpenCL::A, OpenCL::RG, OpenCL::RA, OpenCL::RGB,
+OpenCL::RGBA, OpenCL::BGRA, OpenCL::ARGB, OpenCL::INTENSITY, OpenCL::LUMINANCE,
+OpenCL::Rx, OpenCL::RGx, OpenCL::RGBx.
+
+channel_type: OpenCL::SNORM_INT8, OpenCL::SNORM_INT16, OpenCL::UNORM_INT8,
+OpenCL::UNORM_INT16, OpenCL::UNORM_SHORT_565, OpenCL::UNORM_SHORT_555,
+OpenCL::UNORM_INT_101010, OpenCL::SIGNED_INT8, OpenCL::SIGNED_INT16,
+OpenCL::SIGNED_INT32, OpenCL::UNSIGNED_INT8, OpenCL::UNSIGNED_INT16,
+OpenCL::UNSIGNED_INT32, OpenCL::HALF_FLOAT, OpenCL::FLOAT.
+
+type: OpenCL::MEM_OBJECT_BUFFER, OpenCL::MEM_OBJECT_IMAGE2D,
+OpenCL::MEM_OBJECT_IMAGE3D, OpenCL::MEM_OBJECT_IMAGE2D_ARRAY,
+OpenCL::MEM_OBJECT_IMAGE1D, OpenCL::MEM_OBJECT_IMAGE1D_ARRAY,
+OpenCL::MEM_OBJECT_IMAGE1D_BUFFER.
+
 L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateImage.html>
 
 =item $img = $ctx->image2d ($flags, $channel_order, $channel_type, $width, $height, $row_pitch = 0, $data = undef)
@@ -1111,6 +1266,8 @@
 Creates a new OpenCL::Buffer (actually OpenCL::BufferObj) object that refers to the given
 OpenGL buffer object.
 
+flags: OpenCL::MEM_READ_WRITE, OpenCL::MEM_READ_ONLY, OpenCL::MEM_WRITE_ONLY.
+
 http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLBuffer.html
 
 =item $img = $ctx->gl_texture ($flags, $target, $miplevel, $texture)
@@ -1118,6 +1275,13 @@
 Creates a new OpenCL::Image object that refers to the given OpenGL
 texture object or buffer.
 
+target: GL_TEXTURE_1D, GL_TEXTURE_1D_ARRAY, GL_TEXTURE_BUFFER,
+GL_TEXTURE_2D, GL_TEXTURE_2D_ARRAY, GL_TEXTURE_3D,
+GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_Y,
+GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_X,
+GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, GL_TEXTURE_CUBE_MAP_NEGATIVE_Z,
+GL_TEXTURE_RECTANGLE/GL_TEXTURE_RECTANGLE_ARB.
+
 http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateFromGLTexture.html
 
 =item $img = $ctx->gl_texture2d ($flags, $target, $miplevel, $texture)
@@ -1152,6 +1316,11 @@
 
 Creates a new OpenCL::Sampler object.
 
+addressing_mode: OpenCL::ADDRESS_NONE, OpenCL::ADDRESS_CLAMP_TO_EDGE,
+OpenCL::ADDRESS_CLAMP, OpenCL::ADDRESS_REPEAT, OpenCL::ADDRESS_MIRRORED_REPEAT.
+
+filter_mode: OpenCL::FILTER_NEAREST, OpenCL::FILTER_LINEAR.
+
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateSampler.html>
 
 =item $program = $ctx->program_with_source ($string)
@@ -1171,6 +1340,19 @@
 
    my $clone = $ctx->program_with_binary ([$prog->devices], [$prog->binaries]);
 
+=item $program = $ctx->program_with_built_in_kernels (\@devices, $kernel_names)
+
+Creates a new OpenCL::Program object from the given built-in kernel names.
+
+L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateProgramWithBuiltInKernels.html>
+
+=item $program = $ctx->link_program (\@devices, $options, \@programs, $cb->($program) = undef)
+
+Links all (already compiled) program objects specified in C<@programs>
+together and returns a new OpenCL::Program object with the result.
+
+L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clLinkProgram.html>
+
 =item $packed_value = $ctx->info ($name)
 
 See C<< $platform->info >> for details.
@@ -1181,19 +1363,19 @@
 
 =item $uint = $context->reference_count
 
-Calls C<clGetContextInfo> with C<CL_CONTEXT_REFERENCE_COUNT> and returns the result.
+Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_REFERENCE_COUNT> and returns the result.
 
 =item @devices = $context->devices
 
-Calls C<clGetContextInfo> with C<CL_CONTEXT_DEVICES> and returns the result.
+Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_DEVICES> and returns the result.
 
 =item @property_ints = $context->properties
 
-Calls C<clGetContextInfo> with C<CL_CONTEXT_PROPERTIES> and returns the result.
+Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_PROPERTIES> and returns the result.
 
 =item $uint = $context->num_devices
 
-Calls C<clGetContextInfo> with C<CL_CONTEXT_NUM_DEVICES> and returns the result.
+Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_NUM_DEVICES> and returns the result.
 
 =for gengetinfo end context
 
@@ -1319,6 +1501,14 @@
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueNDRangeKernel.html>
 
+=item $ev = $queue->migrate_mem_objects (\@mem_objects, $flags, $wait_events...)
+
+Migrates a number of OpenCL::Memory objects to or from the device.
+
+flags: OpenCL::MIGRATE_MEM_OBJECT_HOST, OpenCL::MIGRATE_MEM_OBJECT_CONTENT_UNDEFINED
+
+L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueMigrateMemObjects.html>
+
 =item $ev = $queue->acquire_gl_objects ([object, ...], $wait_events...)
 
 Enqueues a list (an array-ref of OpenCL::Memory objects) to be acquired
@@ -1363,19 +1553,19 @@
 
 =item $ctx = $command_queue->context
 
-Calls C<clGetCommandQueueInfo> with C<CL_QUEUE_CONTEXT> and returns the result.
+Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_CONTEXT> and returns the result.
 
 =item $device = $command_queue->device
 
-Calls C<clGetCommandQueueInfo> with C<CL_QUEUE_DEVICE> and returns the result.
+Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_DEVICE> and returns the result.
 
 =item $uint = $command_queue->reference_count
 
-Calls C<clGetCommandQueueInfo> with C<CL_QUEUE_REFERENCE_COUNT> and returns the result.
+Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_REFERENCE_COUNT> and returns the result.
 
 =item $command_queue_properties = $command_queue->properties
 
-Calls C<clGetCommandQueueInfo> with C<CL_QUEUE_PROPERTIES> and returns the result.
+Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_PROPERTIES> and returns the result.
 
 =for gengetinfo end command_queue
 
@@ -1404,6 +1594,8 @@
 C<OpenCL::MappedBuffer> object. If C<$size> is specified as undef, then
 the map will extend to the end of the buffer.
 
+map_flags: OpenCL::MAP_READ, OpenCL::MAP_WRITE, OpenCL::MAP_WRITE_INVALIDATE_REGION.
+
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueMapBuffer.html>
 
 Example: map the buffer $buf fully and replace the first 4 bytes by "abcd", then unmap.
@@ -1465,39 +1657,39 @@
 
 =item $mem_object_type = $mem->type
 
-Calls C<clGetMemObjectInfo> with C<CL_MEM_TYPE> and returns the result.
+Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_TYPE> and returns the result.
 
 =item $mem_flags = $mem->flags
 
-Calls C<clGetMemObjectInfo> with C<CL_MEM_FLAGS> and returns the result.
+Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_FLAGS> and returns the result.
 
 =item $int = $mem->size
 
-Calls C<clGetMemObjectInfo> with C<CL_MEM_SIZE> and returns the result.
+Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_SIZE> and returns the result.
 
 =item $ptr_value = $mem->host_ptr
 
-Calls C<clGetMemObjectInfo> with C<CL_MEM_HOST_PTR> and returns the result.
+Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_HOST_PTR> and returns the result.
 
 =item $uint = $mem->map_count
 
-Calls C<clGetMemObjectInfo> with C<CL_MEM_MAP_COUNT> and returns the result.
+Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_MAP_COUNT> and returns the result.
 
 =item $uint = $mem->reference_count
 
-Calls C<clGetMemObjectInfo> with C<CL_MEM_REFERENCE_COUNT> and returns the result.
+Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_REFERENCE_COUNT> and returns the result.
 
 =item $ctx = $mem->context
 
-Calls C<clGetMemObjectInfo> with C<CL_MEM_CONTEXT> and returns the result.
+Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_CONTEXT> and returns the result.
 
 =item $mem = $mem->associated_memobject
 
-Calls C<clGetMemObjectInfo> with C<CL_MEM_ASSOCIATED_MEMOBJECT> and returns the result.
+Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_ASSOCIATED_MEMOBJECT> and returns the result.
 
 =item $int = $mem->offset
 
-Calls C<clGetMemObjectInfo> with C<CL_MEM_OFFSET> and returns the result.
+Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_OFFSET> and returns the result.
 
 =for gengetinfo end mem
 
@@ -1527,7 +1719,7 @@
 =item $subbuf = $buf_obj->sub_buffer_region ($flags, $origin, $size)
 
 Creates an OpenCL::Buffer objects from this buffer and returns it. The
-C<buffer_create_type> is assumed to be C<CL_BUFFER_CREATE_TYPE_REGION>.
+C<buffer_create_type> is assumed to be C<OpenCL::BUFFER_CREATE_TYPE_REGION>.
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateSubBuffer.html>
 
@@ -1553,33 +1745,33 @@
 =item ($channel_order, $channel_data_type) = $image->format
 
 Returns the channel order and type used to create the image by calling
-C<clGetImageInfo> with C<CL_IMAGE_FORMAT>.
+C<clGetImageInfo> with C<OpenCL::IMAGE_FORMAT>.
 
 =for gengetinfo begin image
 
 =item $int = $image->element_size
 
-Calls C<clGetImageInfo> with C<CL_IMAGE_ELEMENT_SIZE> and returns the result.
+Calls C<clGetImageInfo> with C<OpenCL::IMAGE_ELEMENT_SIZE> and returns the result.
 
 =item $int = $image->row_pitch
 
-Calls C<clGetImageInfo> with C<CL_IMAGE_ROW_PITCH> and returns the result.
+Calls C<clGetImageInfo> with C<OpenCL::IMAGE_ROW_PITCH> and returns the result.
 
 =item $int = $image->slice_pitch
 
-Calls C<clGetImageInfo> with C<CL_IMAGE_SLICE_PITCH> and returns the result.
+Calls C<clGetImageInfo> with C<OpenCL::IMAGE_SLICE_PITCH> and returns the result.
 
 =item $int = $image->width
 
-Calls C<clGetImageInfo> with C<CL_IMAGE_WIDTH> and returns the result.
+Calls C<clGetImageInfo> with C<OpenCL::IMAGE_WIDTH> and returns the result.
 
 =item $int = $image->height
 
-Calls C<clGetImageInfo> with C<CL_IMAGE_HEIGHT> and returns the result.
+Calls C<clGetImageInfo> with C<OpenCL::IMAGE_HEIGHT> and returns the result.
 
 =item $int = $image->depth
 
-Calls C<clGetImageInfo> with C<CL_IMAGE_DEPTH> and returns the result.
+Calls C<clGetImageInfo> with C<OpenCL::IMAGE_DEPTH> and returns the result.
 
 =for gengetinfo end image
 
@@ -1587,11 +1779,11 @@
 
 =item $GLenum = $gl_texture->target
 
-Calls C<clGetGLTextureInfo> with C<CL_GL_TEXTURE_TARGET> and returns the result.
+Calls C<clGetGLTextureInfo> with C<OpenCL::GL_TEXTURE_TARGET> and returns the result.
 
 =item $GLint = $gl_texture->gl_mipmap_level
 
-Calls C<clGetGLTextureInfo> with C<CL_GL_MIPMAP_LEVEL> and returns the result.
+Calls C<clGetGLTextureInfo> with C<OpenCL::GL_MIPMAP_LEVEL> and returns the result.
 
 =for gengetinfo end gl_texture
 
@@ -1611,23 +1803,23 @@
 
 =item $uint = $sampler->reference_count
 
-Calls C<clGetSamplerInfo> with C<CL_SAMPLER_REFERENCE_COUNT> and returns the result.
+Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_REFERENCE_COUNT> and returns the result.
 
 =item $ctx = $sampler->context
 
-Calls C<clGetSamplerInfo> with C<CL_SAMPLER_CONTEXT> and returns the result.
+Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_CONTEXT> and returns the result.
 
 =item $addressing_mode = $sampler->normalized_coords
 
-Calls C<clGetSamplerInfo> with C<CL_SAMPLER_NORMALIZED_COORDS> and returns the result.
+Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_NORMALIZED_COORDS> and returns the result.
 
 =item $filter_mode = $sampler->addressing_mode
 
-Calls C<clGetSamplerInfo> with C<CL_SAMPLER_ADDRESSING_MODE> and returns the result.
+Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_ADDRESSING_MODE> and returns the result.
 
 =item $boolean = $sampler->filter_mode
 
-Calls C<clGetSamplerInfo> with C<CL_SAMPLER_FILTER_MODE> and returns the result.
+Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_FILTER_MODE> and returns the result.
 
 =for gengetinfo end sampler
 
@@ -1655,6 +1847,17 @@
 Some implementations fail with C<OpenCL::INVALID_BINARY> when the
 compilation state is successful but some later stage fails.
 
+options: C<-D name>, C<-D name=definition>, C<-I dir>,
+C<-cl-single-precision-constant>, C<-cl-denorms-are-zero>,
+C<-cl-fp32-correctly-rounded-divide-sqrt>, C<-cl-opt-disable>,
+C<-cl-mad-enable>, C<-cl-no-signed-zeros>, C<-cl-unsafe-math-optimizations>,
+C<-cl-finite-math-only>, C<-cl-fast-relaxed-math>,
+C<-w>, C<-Werror>, C<-cl-std=CL1.1/CL1.2>, C<-cl-kernel-arg-info>,
+C<-create-library>, C<-enable-link-options>.
+
+build_status: OpenCL::BUILD_SUCCESS, OpenCL::BUILD_NONE,
+OpenCL::BUILD_ERROR, OpenCL::BUILD_IN_PROGRESS.
+
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clBuildProgram.html>
 
 =item $program->build_async (\@devices = undef, $options = "", $cb->($program) = undef)
@@ -1662,11 +1865,24 @@
 Similar to C<< ->build >>, except it starts a thread, and never fails (you
 need to check the compilation status form the callback, or by polling).
 
+=item $program->compile (\@devices = undef, $options = "", \%headers = undef, $cb->($program) = undef)
+
+Compiles the given program for the given devices (or all devices if
+undef). If C<$headers> is given, it must be a hashref with include name =>
+OpenCL::Program pairs.
+
+L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCompileProgram.html>
+
 =item $packed_value = $program->build_info ($device, $name)
 
 Similar to C<< $platform->info >>, but returns build info for a previous
 build attempt for the given device.
 
+binary_type: OpenCL::PROGRAM_BINARY_TYPE_NONE,
+OpenCL::PROGRAM_BINARY_TYPE_COMPILED_OBJECT,
+OpenCL::PROGRAM_BINARY_TYPE_LIBRARY,
+OpenCL::PROGRAM_BINARY_TYPE_EXECUTABLE.
+
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetBuildInfo.html>
 
 =item $kernel = $program->kernel ($function_name)
@@ -1686,15 +1902,19 @@
 
 =item $build_status = $program->build_status ($device)
 
-Calls C<clGetProgramBuildInfo> with C<CL_PROGRAM_BUILD_STATUS> and returns the result.
+Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BUILD_STATUS> and returns the result.
 
 =item $string = $program->build_options ($device)
 
-Calls C<clGetProgramBuildInfo> with C<CL_PROGRAM_BUILD_OPTIONS> and returns the result.
+Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BUILD_OPTIONS> and returns the result.
 
 =item $string = $program->build_log ($device)
 
-Calls C<clGetProgramBuildInfo> with C<CL_PROGRAM_BUILD_LOG> and returns the result.
+Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BUILD_LOG> and returns the result.
+
+=item $binary_type = $program->binary_type ($device)
+
+Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BINARY_TYPE> and returns the result.
 
 =for gengetinfo end program_build
 
@@ -1708,27 +1928,27 @@
 
 =item $uint = $program->reference_count
 
-Calls C<clGetProgramInfo> with C<CL_PROGRAM_REFERENCE_COUNT> and returns the result.
+Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_REFERENCE_COUNT> and returns the result.
 
 =item $ctx = $program->context
 
-Calls C<clGetProgramInfo> with C<CL_PROGRAM_CONTEXT> and returns the result.
+Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_CONTEXT> and returns the result.
 
 =item $uint = $program->num_devices
 
-Calls C<clGetProgramInfo> with C<CL_PROGRAM_NUM_DEVICES> and returns the result.
+Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_NUM_DEVICES> and returns the result.
 
 =item @devices = $program->devices
 
-Calls C<clGetProgramInfo> with C<CL_PROGRAM_DEVICES> and returns the result.
+Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_DEVICES> and returns the result.
 
 =item $string = $program->source
 
-Calls C<clGetProgramInfo> with C<CL_PROGRAM_SOURCE> and returns the result.
+Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_SOURCE> and returns the result.
 
 =item @ints = $program->binary_sizes
 
-Calls C<clGetProgramInfo> with C<CL_PROGRAM_BINARY_SIZES> and returns the result.
+Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_BINARY_SIZES> and returns the result.
 
 =for gengetinfo end program
 
@@ -1759,23 +1979,23 @@
 
 =item $string = $kernel->function_name
 
-Calls C<clGetKernelInfo> with C<CL_KERNEL_FUNCTION_NAME> and returns the result.
+Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_FUNCTION_NAME> and returns the result.
 
 =item $uint = $kernel->num_args
 
-Calls C<clGetKernelInfo> with C<CL_KERNEL_NUM_ARGS> and returns the result.
+Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_NUM_ARGS> and returns the result.
 
 =item $uint = $kernel->reference_count
 
-Calls C<clGetKernelInfo> with C<CL_KERNEL_REFERENCE_COUNT> and returns the result.
+Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_REFERENCE_COUNT> and returns the result.
 
 =item $ctx = $kernel->context
 
-Calls C<clGetKernelInfo> with C<CL_KERNEL_CONTEXT> and returns the result.
+Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_CONTEXT> and returns the result.
 
 =item $program = $kernel->program
 
-Calls C<clGetKernelInfo> with C<CL_KERNEL_PROGRAM> and returns the result.
+Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_PROGRAM> and returns the result.
 
 =for gengetinfo end kernel
 
@@ -1783,35 +2003,62 @@
 
 See C<< $platform->info >> for details.
 
-The reason this method is not called C<info> is that there already is an
-C<< ->info >> method.
-
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetKernelWorkGroupInfo.html>
 
 =for gengetinfo begin kernel_work_group
 
 =item $int = $kernel->work_group_size ($device)
 
-Calls C<clGetKernelWorkGroupInfo> with C<CL_KERNEL_WORK_GROUP_SIZE> and returns the result.
+Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_WORK_GROUP_SIZE> and returns the result.
 
 =item @ints = $kernel->compile_work_group_size ($device)
 
-Calls C<clGetKernelWorkGroupInfo> with C<CL_KERNEL_COMPILE_WORK_GROUP_SIZE> and returns the result.
+Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_COMPILE_WORK_GROUP_SIZE> and returns the result.
 
 =item $ulong = $kernel->local_mem_size ($device)
 
-Calls C<clGetKernelWorkGroupInfo> with C<CL_KERNEL_LOCAL_MEM_SIZE> and returns the result.
+Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_LOCAL_MEM_SIZE> and returns the result.
 
 =item $int = $kernel->preferred_work_group_size_multiple ($device)
 
-Calls C<clGetKernelWorkGroupInfo> with C<CL_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE> and returns the result.
+Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE> and returns the result.
 
 =item $ulong = $kernel->private_mem_size ($device)
 
-Calls C<clGetKernelWorkGroupInfo> with C<CL_KERNEL_PRIVATE_MEM_SIZE> and returns the result.
+Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_PRIVATE_MEM_SIZE> and returns the result.
 
 =for gengetinfo end kernel_work_group
 
+=item $packed_value = $kernel->arg_info ($idx, $name)
+
+See C<< $platform->info >> for details.
+
+L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clGetKernelArgInfo.html>
+
+=for gengetinfo begin kernel_arg
+
+=item $kernel_arg_address_qualifier = $kernel->arg_address_qualifier ($idx)
+
+Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_ADDRESS_QUALIFIER> and returns the result.
+
+=item $kernel_arg_access_qualifier = $kernel->arg_access_qualifier ($idx)
+
+Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_ACCESS_QUALIFIER> and returns the result.
+
+=item $string = $kernel->arg_type_name ($idx)
+
+Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_TYPE_NAME> and returns the result.
+
+=item $kernel_arg_type_qualifier = $kernel->arg_type_qualifier ($idx)
+
+Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_TYPE_QUALIFIER> and returns the result.
+
+=item $string = $kernel->arg_name ($idx)
+
+Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_NAME> and returns the result.
+
+=for gengetinfo end kernel_arg
+
 =item $kernel->setf ($format, ...)
 
 Sets the arguments of a kernel. Since OpenCL 1.1 doesn't have a generic
@@ -1932,23 +2179,23 @@
 
 =item $queue = $event->command_queue
 
-Calls C<clGetEventInfo> with C<CL_EVENT_COMMAND_QUEUE> and returns the result.
+Calls C<clGetEventInfo> with C<OpenCL::EVENT_COMMAND_QUEUE> and returns the result.
 
 =item $command_type = $event->command_type
 
-Calls C<clGetEventInfo> with C<CL_EVENT_COMMAND_TYPE> and returns the result.
+Calls C<clGetEventInfo> with C<OpenCL::EVENT_COMMAND_TYPE> and returns the result.
 
 =item $uint = $event->reference_count
 
-Calls C<clGetEventInfo> with C<CL_EVENT_REFERENCE_COUNT> and returns the result.
+Calls C<clGetEventInfo> with C<OpenCL::EVENT_REFERENCE_COUNT> and returns the result.
 
 =item $uint = $event->command_execution_status
 
-Calls C<clGetEventInfo> with C<CL_EVENT_COMMAND_EXECUTION_STATUS> and returns the result.
+Calls C<clGetEventInfo> with C<OpenCL::EVENT_COMMAND_EXECUTION_STATUS> and returns the result.
 
 =item $ctx = $event->context
 
-Calls C<clGetEventInfo> with C<CL_EVENT_CONTEXT> and returns the result.
+Calls C<clGetEventInfo> with C<OpenCL::EVENT_CONTEXT> and returns the result.
 
 =for gengetinfo end event
 
@@ -1965,19 +2212,19 @@
 
 =item $ulong = $event->profiling_command_queued
 
-Calls C<clGetEventProfilingInfo> with C<CL_PROFILING_COMMAND_QUEUED> and returns the result.
+Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_QUEUED> and returns the result.
 
 =item $ulong = $event->profiling_command_submit
 
-Calls C<clGetEventProfilingInfo> with C<CL_PROFILING_COMMAND_SUBMIT> and returns the result.
+Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_SUBMIT> and returns the result.
 
 =item $ulong = $event->profiling_command_start
 
-Calls C<clGetEventProfilingInfo> with C<CL_PROFILING_COMMAND_START> and returns the result.
+Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_START> and returns the result.
 
 =item $ulong = $event->profiling_command_end
 
-Calls C<clGetEventProfilingInfo> with C<CL_PROFILING_COMMAND_END> and returns the result.
+Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_END> and returns the result.
 
 =for gengetinfo end profiling
 
@@ -1994,6 +2241,8 @@
 Sets the execution status of the user event. Can only be called once,
 either with OpenCL::COMPLETE or a negative number as status.
 
+execution_status: OpenCL::COMPLETE or a negative integer.
+
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetUserEventStatus.html>
 
 =back