--- OpenCL/OpenCL.pm	2011/11/17 02:10:39	1.10
+++ OpenCL/OpenCL.pm	2011/11/20 22:31:48	1.24
@@ -20,18 +20,18 @@
 Each platform gives you access to a number of OpenCL::Device objects, e.g.
 your graphics card.
 
-From a platform and some devices, you create an OpenCL::Context, which is
+From a platform and some device(s), you create an OpenCL::Context, which is
 a very central object in OpenCL: Once you have a context you can create
 most other objects:
 
-OpenCL::Program objects, which store source code and, after building
-("compiling and linking"), also binary programs. For each kernel function
-in a program you can then create an OpenCL::Kernel object which represents
-basically a function call with argument values.
-
-OpenCL::Memory objects of various flavours: OpenCL::Buffers objects (flat
-memory areas, think array) and OpenCL::Image objects (think 2d or 3d
-array) for bulk data and input and output for kernels.
+OpenCL::Program objects, which store source code and, after building for a
+specific device ("compiling and linking"), also binary programs. For each
+kernel function in a program you can then create an OpenCL::Kernel object
+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.
 
 OpenCL::Sampler objects, which are kind of like texture filter modes in
 OpenGL.
@@ -54,38 +54,59 @@
 
    http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/
 
+If you are into UML class diagrams, the following diagram might help - if
+not, it will be mildly cobfusing:
+
+   http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/classDiagram.html
+
+Here's a tutorial from AMD (very AMD-centric, too), not sure how useful it
+is, but at least it's free of charge:
+
+   http://developer.amd.com/zones/OpenCLZone/courses/Documents/Introduction_to_OpenCL_Programming%20Training_Guide%20%28201005%29.pdf
+
+And here's NVIDIA's OpenCL Best Practises Guide:
+
+   http://developer.download.nvidia.com/compute/cuda/3_2/toolkit/docs/OpenCL_Best_Practices_Guide.pdf
+
 =head1 BASIC WORKFLOW
 
-To get something done, you basically have to do this once:
+To get something done, you basically have to do this once (refer to the
+examples below for actual code, this is just a high-level description):
 
-Find some platform (e.g. the first one) and some device (e.g. the first
-device you can find), and create a context from those.
+Find some platform (e.g. the first one) and some device(s) (e.g. the first
+device of the platform), and create a context from those.
 
-Create a command queue from your context, and program objects from your
-OpenCL source code, build the programs.
+Create program objects from your OpenCL source code, then build (compile)
+the programs for each device you want to run them on.
 
-Create kernel objects for all kernels you want to use.
+Create kernel objects for all kernels you want to use (surprisingly, these
+are not device-specific).
 
-Then, to execute stuff, you repeat this:
+Then, to execute stuff, you repeat these steps, possibly resuing or
+sharing some buffers:
 
-Create some input and output buffers from your context. Initialise the
-input buffers with data. Set these as arguments to your kernel.
+Create some input and output buffers from your context. Set these as
+arguments to your kernel.
+
+Enqueue buffer writes to initialise your input buffers (when not
+initialised at creation time).
 
 Enqueue the kernel execution.
 
 Enqueue buffer reads for your output buffer to read results.
 
-The next section shows how this can be done.
-
 =head1 EXAMPLES
 
 =head2 Enumerate all devices and get contexts for them.
 
+Best run this once to get a feel for the platforms and devices in your
+system.
+
    for my $platform (OpenCL::platforms) {
-      printf "platform: %s\n", $platform->info (OpenCL::PLATFORM_NAME);
-      printf "extensions: %s\n", $platform->info (OpenCL::PLATFORM_EXTENSIONS);
+      printf "platform: %s\n", $platform->name;
+      printf "extensions: %s\n", $platform->extensions;
       for my $device ($platform->devices) {
-         printf "+ device: %s\n", $device->info (OpenCL::DEVICE_NAME);
+         printf "+ device: %s\n", $device->name;
          my $ctx = $device->context;
          # do stuff
       }
@@ -93,12 +114,19 @@
 
 =head2 Get a useful context and a command queue.
 
-   my $dev = ((OpenCL::platforms)[0]->devices)[0];
-   my $ctx = $dev->context;
-   my $queue = $ctx->queue ($dev);
+This is a useful boilerplate for any OpenCL program that only wants to use
+one device,
+
+   my ($platform) = OpenCL::platforms; # find first platform
+   my ($dev) = $platform->devices;     # find first device of platform
+   my $ctx = $platform->context (undef, [$dev]); # create context out of those
+   my $queue = $ctx->queue ($dev);     # create a command queue for the device
 
 =head2 Print all supported image formats of a context.
 
+Best run this once for your context, to see whats available and how to
+gather information.
+
    for my $type (OpenCL::MEM_OBJECT_IMAGE2D, OpenCL::MEM_OBJECT_IMAGE3D) {
       print "supported image formats for ", OpenCL::enum2str $type, "\n";
       
@@ -126,19 +154,21 @@
       __kernel void
       squareit (__global float *input, __global float *output)
       {
-        size_t id = get_global_id (0);
+        $id = get_global_id (0);
         output [id] = input [id] * input [id];
       }
    ';
 
    my $prog = $ctx->program_with_source ($src);
 
+   # build croaks on compile errors, so catch it and print the compile errors
    eval { $prog->build ($dev); 1 }
-      or die $prog->build_info ($dev, OpenCL::PROGRAM_BUILD_LOG);
+      or die $prog->build_log;
 
    my $kernel = $prog->kernel ("squareit");
 
-=head2 Create some input and output float buffers, then call squareit on them.
+=head2 Create some input and output float buffers, then call the
+'squareit' kernel on them.
 
    my $input  = $ctx->buffer_sv (OpenCL::MEM_COPY_HOST_PTR, pack "f*", 1, 2, 3, 4.5);
    my $output = $ctx->buffer (0, OpenCL::SIZEOF_FLOAT * 5);
@@ -187,8 +217,8 @@
 
 =head2 BASIC CONVENTIONS
 
-This is not a 1:1 C-style translation of OpenCL to Perl - instead I
-attempted to make the interface as type-safe as possible and introducing
+This is not a one-to-one C-style translation of OpenCL to Perl - instead
+I attempted to make the interface as type-safe as possible by introducing
 object syntax where it makes sense. There are a number of important
 differences between the OpenCL C API and this module:
 
@@ -198,16 +228,18 @@
 to free objects explicitly (C<clReleaseXXX>), the release function
 is called automatically once all Perl references to it go away.
 
-=item * OpenCL uses CamelCase for function names (C<clGetPlatformInfo>),
-while this module uses underscores as word separator and often leaves out
-prefixes (C<< $platform->info >>).
+=item * OpenCL uses CamelCase for function names
+(e.g. C<clGetPlatformIDs>, C<clGetPlatformInfo>), while this module
+uses underscores as word separator and often leaves out prefixes
+(C<OpenCL::platforms>, C<< $platform->info >>).
 
 =item * OpenCL often specifies fixed vector function arguments as short
 arrays (C<size_t origin[3]>), while this module explicitly expects the
-components as separate arguments-
+components as separate arguments (C<$orig_x, $orig_y, $orig_z>) in
+function calls.
 
-=item * Where possible, one of the pitch values is calculated from the
-perl scalar length and need not be specified.
+=item * Structures are often specified by flattening out their components
+as with short vectors, and returned as arrayrefs.
 
 =item * When enqueuing commands, the wait list is specified by adding
 extra arguments to the function - anywhere a C<$wait_events...> argument
@@ -247,7 +279,8 @@
 
 =item $int = OpenCL::errno
 
-The last error returned by a function - it's only changed on errors.
+The last error returned by a function - it's only valid after an error occured
+and before calling another OpenCL function.
 
 =item $str = OpenCL::err2str $errval
 
@@ -285,23 +318,57 @@
 
 =over 4
 
+=item @devices = $platform->devices ($type = OpenCL::DEVICE_TYPE_ALL)
+
+Returns a list of matching OpenCL::Device objects.
+
+=item $ctx = $platform->context_from_type ($properties, $type = OpenCL::DEVICE_TYPE_DEFAULT, $notify = undef)
+
+Tries to create a context. Never worked for me, and you need devices explicitly anyway.
+
+L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContextFromType.html>
+
+=item $ctx = $device->context ($properties = undef, @$devices, $notify = undef)
+
+Create a new OpenCL::Context object using the given device object(s)- a
+CL_CONTEXT_PLATFORM property is supplied automatically.
+
+L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContext.html>
+
 =item $packed_value = $platform->info ($name)
 
 Calls C<clGetPlatformInfo> and returns the packed, raw value - for
-strings, this will be the string, for other values you probably need to
-use the correct C<unpack>. This might get improved in the future. Hopefully.
+strings, this will be the string (possibly including terminating \0), for
+other values you probably need to use the correct C<unpack>.
+
+It's best to avoid this method and use one of the following convenience
+wrappers.
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetPlatformInfo.html>
 
-=item @devices = $platform->devices ($type = OpenCL::DEVICE_TYPE_ALL)
+=for gengetinfo begin platform
 
-Returns a list of matching OpenCL::Device objects.
+=item $string = $platform->profile
 
-=item $ctx = $platform->context_from_type ($properties, $type = OpenCL::DEVICE_TYPE_DEFAULT, $notify = undef)
+Calls C<clGetPlatformInfo> with C<CL_PLATFORM_PROFILE> and returns the result.
 
-Tries to create a context. Never worked for me.
+=item $string = $platform->version
 
-L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContextFromType.html>
+Calls C<clGetPlatformInfo> with C<CL_PLATFORM_VERSION> and returns the result.
+
+=item $string = $platform->name
+
+Calls C<clGetPlatformInfo> with C<CL_PLATFORM_NAME> and returns the result.
+
+=item $string = $platform->vendor
+
+Calls C<clGetPlatformInfo> with C<CL_PLATFORM_VENDOR> and returns the result.
+
+=item $string = $platform->extensions
+
+Calls C<clGetPlatformInfo> with C<CL_PLATFORM_EXTENSIONS> and returns the result.
+
+=for gengetinfo end platform
 
 =back
 
@@ -315,23 +382,279 @@
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetDeviceInfo.html>
 
-=item $ctx = $device->context ($properties = undef, $notify = undef)
+=for gengetinfo begin device
 
-Create a new OpenCL::Context object.
+=item $device_type = $device->type
 
-L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContext.html>
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_TYPE> and returns the result.
 
-=back
+=item $uint = $device->vendor_id
 
-=head2 THE OpenCL::Context CLASS
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_VENDOR_ID> and returns the result.
 
-=over 4
+=item $uint = $device->max_compute_units
 
-=item $packed_value = $ctx->info ($name)
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_MAX_COMPUTE_UNITS> and returns the result.
 
-See C<< $platform->info >> for details.
+=item $uint = $device->max_work_item_dimensions
 
-L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetContextInfo.html>
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item @ints = $device->max_work_item_sizes
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->preferred_vector_width_short
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->preferred_vector_width_long
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->preferred_vector_width_double
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $bitfield = $device->address_bits
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->max_write_image_args
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $int = $device->image2d_max_width
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $int = $device->image3d_max_width
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $int = $device->image3d_max_depth
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $int = $device->max_parameter_size
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->mem_base_addr_align
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $device_fp_config = $device->single_fp_config
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->global_mem_cacheline_size
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $ulong = $device->global_mem_size
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->max_constant_args
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $ulong = $device->local_mem_size
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $int = $device->profiling_timer_resolution
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $boolean = $device->available
+
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_AVAILABLE> and returns the result.
+
+=item $boolean = $device->compiler_available
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $command_queue_properties = $device->properties
+
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_QUEUE_PROPERTIES> and returns the result.
+
+=item $ = $device->platform
+
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_PLATFORM> and returns the result.
+
+=item $string = $device->name
+
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_NAME> and returns the result.
+
+=item $string = $device->vendor
+
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_VENDOR> and returns the result.
+
+=item $string = $device->driver_version
+
+Calls C<clGetDeviceInfo> with C<CL_DRIVER_VERSION> and returns the result.
+
+=item $string = $device->profile
+
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_PROFILE> and returns the result.
+
+=item $string = $device->version
+
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_VERSION> and returns the result.
+
+=item $string = $device->extensions
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->native_vector_width_char
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->native_vector_width_int
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->native_vector_width_float
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $uint = $device->native_vector_width_half
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $device_fp_config = $device->half_fp_config
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item $device = $device->parent_device_ext
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item @device_partition_property_exts = $device->affinity_domains_ext
+
+Calls C<clGetDeviceInfo> with C<CL_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.
+
+=item @device_partition_property_exts = $device->partition_style_ext
+
+Calls C<clGetDeviceInfo> with C<CL_DEVICE_PARTITION_STYLE_EXT> and returns the result.
+
+=for gengetinfo end device
+
+=back
+
+=head2 THE OpenCL::Context CLASS
+
+=over 4
 
 =item $queue = $ctx->queue ($device, $properties)
 
@@ -355,13 +678,13 @@
 
 Creates a new OpenCL::Buffer object and initialise it with the given data values.
 
-=item $img = $ctx->image2d ($flags, $channel_order, $channel_type, $width, $height, $data)
+=item $img = $ctx->image2d ($flags, $channel_order, $channel_type, $width, $height, $row_pitch = 0, $data = undef)
 
 Creates a new OpenCL::Image2D object and optionally initialises it with the given data values.
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateImage2D.html>
 
-=item $img = $ctx->image3d ($flags, $channel_order, $channel_type, $width, $height, $depth, $slice_pitch, $data)
+=item $img = $ctx->image3d ($flags, $channel_order, $channel_type, $width, $height, $depth, $row_pitch = 0, $slice_pitch = 0, $data = undef)
 
 Creates a new OpenCL::Image3D object and optionally initialises it with the given data values.
 
@@ -386,6 +709,32 @@
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateProgramWithSource.html>
 
+=item $packed_value = $ctx->info ($name)
+
+See C<< $platform->info >> for details.
+
+L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetContextInfo.html>
+
+=for gengetinfo begin context
+
+=item $uint = $context->reference_count
+
+Calls C<clGetContextInfo> with C<CL_CONTEXT_REFERENCE_COUNT> and returns the result.
+
+=item @devices = $context->devices
+
+Calls C<clGetContextInfo> with C<CL_CONTEXT_DEVICES> and returns the result.
+
+=item @property_ints = $context->properties
+
+Calls C<clGetContextInfo> with C<CL_CONTEXT_PROPERTIES> and returns the result.
+
+=item $uint = $context->num_devices
+
+Calls C<clGetContextInfo> with C<CL_CONTEXT_NUM_DEVICES> and returns the result.
+
+=for gengetinfo end context
+
 =back
 
 =head2 THE OpenCL::Queue CLASS
@@ -408,12 +757,6 @@
 
 =over 4
 
-=item $packed_value = $ctx->info ($name)
-
-See C<< $platform->info >> for details.
-
-L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetCommandQueueInfo.html>
-
 =item $ev = $queue->enqueue_read_buffer ($buffer, $blocking, $offset, $len, $data, $wait_events...)
 
 Reads data from buffer into the given string.
@@ -434,25 +777,25 @@
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadImage.html>
 
-=item $ev = $queue->enqueue_write_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $data, $wait_events...)
+=item $ev = $queue->enqueue_write_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $slice_pitch, $data, $wait_events...)
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteImage.html>
 
-=item $ev = $queue->enqueue_copy_buffer_rect ($src, $dst, $src_x, $src_y, $src_z, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $src_row_pitch, $src_slice_pitch, 4dst_row_pitch, $dst_slice_pitch, $ait_event...)
+=item $ev = $queue->enqueue_copy_buffer_rect ($src, $dst, $src_x, $src_y, $src_z, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $src_row_pitch, $src_slice_pitch, $dst_row_pitch, $dst_slice_pitch, $wait_event...)
 
 Yeah.
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBufferRect.html>
 
-=item $ev = $queue->enqueue_copy_buffer_to_image (OpenCL::Buffer src, OpenCL::Image dst, size_t src_offset, size_t dst_x, size_t dst_y, size_t dst_z, size_t width, size_t height, size_t depth, ...)
+=item $ev = $queue->enqueue_copy_buffer_to_image ($src_buffer, $dst_image, $src_offset, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $wait_events...)
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBufferToImage.html>.
 
-=item $ev = $queue->enqueue_copy_image (OpenCL::Image src, OpenCL::Buffer dst, size_t src_x, size_t src_y, size_t src_z, size_t dst_x, size_t dst_y, size_t dst_z, size_t width, size_t height, size_t depth, ...)
+=item $ev = $queue->enqueue_copy_image ($src_image, $dst_image, $src_x, $src_y, $src_z, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $wait_events...)
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyImage.html>
 
-=item $ev = $queue->enqueue_copy_image_to_buffer (OpenCL::Image src, OpenCL::Buffer dst, size_t src_x, size_t src_y, size_t src_z, size_t width, size_t height, size_t depth, size_t dst_offset, ...)
+=item $ev = $queue->enqueue_copy_image_to_buffer ($src_image, $dst_image, $src_x, $src_y, $src_z, $width, $height, $depth, $dst_offset, $wait_events...)
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyImageToBuffer.html>
 
@@ -498,13 +841,38 @@
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFinish.html>
 
+=item $packed_value = $queue->info ($name)
+
+See C<< $platform->info >> for details.
+
+L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetCommandQueueInfo.html>
+
+=for gengetinfo begin command_queue
+
+=item $ctx = $command_queue->context
+
+Calls C<clGetCommandQueueInfo> with C<CL_QUEUE_CONTEXT> and returns the result.
+
+=item $device = $command_queue->device
+
+Calls C<clGetCommandQueueInfo> with C<CL_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.
+
+=item $command_queue_properties = $command_queue->properties
+
+Calls C<clGetCommandQueueInfo> with C<CL_QUEUE_PROPERTIES> and returns the result.
+
+=for gengetinfo end command_queue
+
 =back
 
 =head2 THE OpenCL::Memory CLASS
 
 This the superclass of all memory objects - OpenCL::Buffer, OpenCL::Image,
-OpenCL::Image2D and OpenCL::Image3D. The subclasses of this class
-currently only exist to allow type-checking.
+OpenCL::Image2D and OpenCL::Image3D.
 
 =over 4
 
@@ -514,29 +882,132 @@
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetMemObjectInfo.html>
 
+=for gengetinfo begin mem
+
+=item $mem_object_type = $mem->type
+
+Calls C<clGetMemObjectInfo> with C<CL_MEM_TYPE> and returns the result.
+
+=item $mem_flags = $mem->flags
+
+Calls C<clGetMemObjectInfo> with C<CL_MEM_FLAGS> and returns the result.
+
+=item $int = $mem->size
+
+Calls C<clGetMemObjectInfo> with C<CL_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.
+
+=item $uint = $mem->map_count
+
+Calls C<clGetMemObjectInfo> with C<CL_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.
+
+=item $ctx = $mem->context
+
+Calls C<clGetMemObjectInfo> with C<CL_MEM_CONTEXT> and returns the result.
+
+=item $mem = $mem->associated_memobject
+
+Calls C<clGetMemObjectInfo> with C<CL_MEM_ASSOCIATED_MEMOBJECT> and returns the result.
+
+=item $int = $mem->offset
+
+Calls C<clGetMemObjectInfo> with C<CL_MEM_OFFSET> and returns the result.
+
+=for gengetinfo end mem
+
 =back
 
-=head2 THE OpenCL::Sampler CLASS
+=head2 THE OpenCL::Image CLASS
+
+This is the superclass of all image objects - OpenCL::Image2D and OpenCL::Image3D.
 
 =over 4
 
-=item $packed_value = $sampler->info ($name)
+=item $packed_value = $ev->image_info ($name)
 
 See C<< $platform->info >> for details.
 
-L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetSamplerInfo.html>
+The reason this method is not called C<info> is that there already is an
+C<< ->info >> method inherited from C<OpenCL::Memory>.
+
+L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetImageInfo.html>
+
+=for gengetinfo begin image
+
+=item $int = $image->element_size
+
+Calls C<clGetImageInfo> with C<CL_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.
+
+=item $int = $image->slice_pitch
+
+Calls C<clGetImageInfo> with C<CL_IMAGE_SLICE_PITCH> and returns the result.
+
+=item $int = $image->width
+
+Calls C<clGetImageInfo> with C<CL_IMAGE_WIDTH> and returns the result.
+
+=item $int = $image->height
+
+Calls C<clGetImageInfo> with C<CL_IMAGE_HEIGHT> and returns the result.
+
+=item $int = $image->depth
+
+Calls C<clGetImageInfo> with C<CL_IMAGE_DEPTH> and returns the result.
+
+=for gengetinfo end image
 
 =back
 
-=head2 THE OpenCL::Program CLASS
+=head2 THE OpenCL::Sampler CLASS
 
 =over 4
 
-=item $packed_value = $program->info ($name)
+=item $packed_value = $sampler->info ($name)
 
 See C<< $platform->info >> for details.
 
-L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProgramInfo.html>
+L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetSamplerInfo.html>
+
+=for gengetinfo begin sampler
+
+=item $uint = $sampler->reference_count
+
+Calls C<clGetSamplerInfo> with C<CL_SAMPLER_REFERENCE_COUNT> and returns the result.
+
+=item $ctx = $sampler->context
+
+Calls C<clGetSamplerInfo> with C<CL_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.
+
+=item $filter_mode = $sampler->addressing_mode
+
+Calls C<clGetSamplerInfo> with C<CL_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.
+
+=for gengetinfo end sampler
+
+=back
+
+=head2 THE OpenCL::Program CLASS
+
+=over 4
 
 =item $program->build ($device, $options = "")
 
@@ -558,6 +1029,67 @@
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateKernel.html>
 
+=for gengetinfo begin program_build
+
+=item $build_status = $program->build_status ($device)
+
+Calls C<clGetProgramBuildInfo> with C<CL_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.
+
+=item $string = $program->build_log ($device)
+
+Calls C<clGetProgramBuildInfo> with C<CL_PROGRAM_BUILD_LOG> and returns the result.
+
+=for gengetinfo end program_build
+
+=item $packed_value = $program->info ($name)
+
+See C<< $platform->info >> for details.
+
+L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProgramInfo.html>
+
+=for gengetinfo begin program
+
+=item $uint = $program->reference_count
+
+Calls C<clGetProgramInfo> with C<CL_PROGRAM_REFERENCE_COUNT> and returns the result.
+
+=item $ctx = $program->context
+
+Calls C<clGetProgramInfo> with C<CL_PROGRAM_CONTEXT> and returns the result.
+
+=item $uint = $program->num_devices
+
+Calls C<clGetProgramInfo> with C<CL_PROGRAM_NUM_DEVICES> and returns the result.
+
+=item @devices = $program->devices
+
+Calls C<clGetProgramInfo> with C<CL_PROGRAM_DEVICES> and returns the result.
+
+=item $string = $program->source
+
+Calls C<clGetProgramInfo> with C<CL_PROGRAM_SOURCE> and returns the result.
+
+=item @ints = $program->binary_sizes
+
+Calls C<clGetProgramInfo> with C<CL_PROGRAM_BINARY_SIZES> and returns the result.
+
+=for gengetinfo end program
+
+=item @blobs = $program->binaries
+
+Returns a string for the compiled binary for every device associated with
+the program, empty strings indicate missing programs, and an empty result
+means no program binaries are available.
+
+These "binaries" are often, in fact, informative low-level assembly
+sources.
+
+L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProgramInfo.html>
+
 =back
 
 =head2 THE OpenCL::Kernel CLASS
@@ -570,6 +1102,63 @@
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetKernelInfo.html>
 
+=for gengetinfo begin kernel
+
+=item $string = $kernel->function_name
+
+Calls C<clGetKernelInfo> with C<CL_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.
+
+=item $uint = $kernel->reference_count
+
+Calls C<clGetKernelInfo> with C<CL_KERNEL_REFERENCE_COUNT> and returns the result.
+
+=item $ctx = $kernel->context
+
+Calls C<clGetKernelInfo> with C<CL_KERNEL_CONTEXT> and returns the result.
+
+=item $program = $kernel->program
+
+Calls C<clGetKernelInfo> with C<CL_KERNEL_PROGRAM> and returns the result.
+
+=for gengetinfo end kernel
+
+=item $packed_value = $kernel->work_group_info ($device, $name)
+
+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.
+
+=item @ints = $kernel->compile_work_group_size ($device)
+
+Calls C<clGetKernelWorkGroupInfo> with C<CL_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.
+
+=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.
+
+=item $ulong = $kernel->private_mem_size ($device)
+
+Calls C<clGetKernelWorkGroupInfo> with C<CL_KERNEL_PRIVATE_MEM_SIZE> and returns the result.
+
+=for gengetinfo end kernel_work_group
+
 =item $kernel->set_TYPE ($index, $value)
 
 This is a family of methods to set the kernel argument with the number C<$index> to the give C<$value>.
@@ -594,17 +1183,70 @@
 
 =over 4
 
+=item $ev->wait
+
+Waits for the event to complete.
+
+L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clWaitForEvents.html>
+
 =item $packed_value = $ev->info ($name)
 
 See C<< $platform->info >> for details.
 
 L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetEventInfo.html>
 
-=item $ev->wait
+=for gengetinfo begin event
 
-Waits for the event to complete.
+=item $queue = $event->command_queue
 
-L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clWaitForEvents.html>
+Calls C<clGetEventInfo> with C<CL_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.
+
+=item $uint = $event->reference_count
+
+Calls C<clGetEventInfo> with C<CL_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.
+
+=item $ctx = $event->context
+
+Calls C<clGetEventInfo> with C<CL_EVENT_CONTEXT> and returns the result.
+
+=for gengetinfo end event
+
+=item $packed_value = $ev->profiling_info ($name)
+
+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/clGetProfilingInfo.html>
+
+=for gengetinfo begin profiling
+
+=item $ulong = $event->profiling_command_queued
+
+Calls C<clGetEventProfilingInfo> with C<CL_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.
+
+=item $ulong = $event->profiling_command_start
+
+Calls C<clGetEventProfilingInfo> with C<CL_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.
+
+=for gengetinfo end profiling
 
 =back
 
@@ -627,7 +1269,7 @@
 use common::sense;
 
 BEGIN {
-   our $VERSION = '0.03';
+   our $VERSION = '0.55';
 
    require XSLoader;
    XSLoader::load (__PACKAGE__, $VERSION);