--- OpenCL/OpenCL.pm 2011/12/10 23:07:38 1.28 +++ OpenCL/OpenCL.pm 2012/04/19 19:36:24 1.38 @@ -107,7 +107,7 @@ printf "extensions: %s\n", $platform->extensions; for my $device ($platform->devices) { printf "+ device: %s\n", $device->name; - my $ctx = $device->context; + my $ctx = $platform->context (undef, [$device]); # do stuff } } @@ -151,8 +151,8 @@ functions. my $src = ' - __kernel void - squareit (__global float *input, __global float *output) + kernel void + squareit (global float *input, global float *output) { $id = get_global_id (0); output [id] = input [id] * input [id]; @@ -213,6 +213,101 @@ # wait for the last event to complete $ev->wait; +=head2 Use the OpenGL module to share a texture between OpenCL and OpenGL and draw some julia +set tunnel effect. + +This is quite a long example to get you going. + + use OpenGL ":all"; + use OpenCL; + + # open a window and create a gl texture + OpenGL::glpOpenWindow width => 256, height => 256; + my $texid = glGenTextures_p 1; + glBindTexture GL_TEXTURE_2D, $texid; + glTexImage2D_c GL_TEXTURE_2D, 0, GL_RGBA8, 256, 256, 0, GL_RGBA, GL_UNSIGNED_BYTE, 0; + + # find and use the first opencl device that let's us get a shared opengl context + my $platform; + my $dev; + my $ctx; + + for (OpenCL::platforms) { + $platform = $_; + for ($platform->devices) { + $dev = $_; + $ctx = $platform->context ([OpenCL::GLX_DISPLAY_KHR, undef, OpenCL::GL_CONTEXT_KHR, undef], [$dev]) + and last; + } + } + + $ctx + or die "cannot find suitable OpenCL device\n"; + + my $queue = $ctx->queue ($dev); + + # now attach an opencl image2d object to the opengl texture + my $tex = $ctx->gl_texture2d (OpenCL::MEM_WRITE_ONLY, GL_TEXTURE_2D, 0, $texid); + + # now the boring opencl code + my $src = <program_with_source ($src); + eval { $prog->build ($dev); 1 } + or die $prog->build_log ($dev); + + my $kernel = $prog->kernel ("juliatunnel"); + + # program compiled, kernel ready, now draw and loop + + for (my $time; ; ++$time) { + # acquire objects from opengl + $queue->enqueue_acquire_gl_objects ([$tex]); + + # configure and run our kernel + $kernel->set_image2d (0, $tex); + $kernel->set_float (1, $time); + $queue->enqueue_nd_range_kernel ($kernel, undef, [256, 256], undef); + + # release objects to opengl again + $queue->enqueue_release_gl_objects ([$tex]); + + # wait + $queue->flush; + + # now draw the texture, the defaults should be all right + glTexParameterf GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST; + + glEnable GL_TEXTURE_2D; + glBegin GL_QUADS; + glTexCoord2f 0, 1; glVertex3i -1, -1, -1; + glTexCoord2f 0, 0; glVertex3i 1, -1, -1; + glTexCoord2f 1, 0; glVertex3i 1, 1, -1; + glTexCoord2f 1, 1; glVertex3i -1, 1, -1; + glEnd; + + glXSwapBuffers; + + select undef, undef, undef, 1/60; + } + =head1 DOCUMENTATION =head2 BASIC CONVENTIONS @@ -273,6 +368,21 @@ half IV ushort S double NV double d +=head2 GLX SUPPORT + +Due to the sad state that OpenGL support is in in Perl (mostly the OpenGL +module, which has little to no documentation and has little to no support +for glX), this module, as a special extension, treats context creation +properties C and C +specially: If either or both of these are C, then the OpenCL +module tries to dynamically resolve C and +C, call these functions and use their return values +instead. + +For this to work, the OpenGL library must be loaded, a GLX context must +have been created and be made current, and C must be available and +capable of finding the function via C. + =head2 THE OpenCL PACKAGE =over 4 @@ -288,9 +398,9 @@ =item $str = OpenCL::enum2str $enum -Converts most enum values (inof parameter names, image format constants, +Converts most enum values (of parameter names, image format constants, object types, addressing and filter modes, command types etc.) into a -human readbale string. When confronted with some random integer it can be +human readable string. When confronted with some random integer it can be very helpful to pass it through this function to maybe get some readable string out of it. @@ -328,7 +438,7 @@ L -=item $ctx = $device->context ($properties = undef, @$devices, $notify = undef) +=item $ctx = $platform->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. @@ -694,6 +804,34 @@ L +=item $buffer = $ctx->gl_buffer ($flags, $bufobj) + +Creates a new OpenCL::Buffer (actually OpenCL::BufferObj) object that refers to the given +OpenGL buffer object. + +http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLBuffer.html + +=item $ctx->gl_texture2d ($flags, $target, $miplevel, $texture) + +Creates a new OpenCL::Image2D object that refers to the given OpenGL +2D texture object. + +http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLTexture2D.html + +=item $ctx->gl_texture3d ($flags, $target, $miplevel, $texture) + +Creates a new OpenCL::Image3D object that refers to the given OpenGL +3D texture object. + +http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLTexture3D.html + +=item $ctx->gl_renderbuffer ($flags, $renderbuffer) + +Creates a new OpenCL::Image2D object that refers to the given OpenGL +render buffer. + +http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLRenderbuffer.html + =item @formats = $ctx->supported_image_formats ($flags, $image_type) Returns a list of matching image formats - each format is an arrayref with @@ -833,10 +971,24 @@ L -=item $ev = $queue->enqueue_marker +=item $ev = $queue->enqueue_marker ($wait_events...) L +=item $ev = $queue->enqueue_acquire_gl_objects ([object, ...], $wait_events...) + +Enqueues a list (an array-ref of OpenCL::Memory objects) to be acquired +for subsequent OpenCL usage. + +L + +=item $ev = $queue->enqueue_release_gl_objects ([object, ...], $wait_events...) + +Enqueues a list (an array-ref of OpenCL::Memory objects) to be released +for subsequent OpenGL usage. + +L + =item $ev = $queue->enqueue_wait_for_events ($wait_events...) L @@ -934,6 +1086,13 @@ =for gengetinfo end mem +=item ($type, $name) = $mem->gl_object_info + +Returns the OpenGL object type (e.g. OpenCL::GL_OBJECT_TEXTURE2D) and the +object "name" (e.g. the texture name) used to create this memory object. + +L + =back =head2 THE OpenCL::Buffer CLASS @@ -1002,6 +1161,18 @@ =for gengetinfo end image +=for gengetinfo begin gl_texture + +=item $GLenum = $gl_texture->target + +Calls C with C and returns the result. + +=item $GLint = $gl_texture->gl_mipmap_level + +Calls C with C and returns the result. + +=for gengetinfo end gl_texture + =back =head2 THE OpenCL::Sampler CLASS @@ -1304,7 +1475,7 @@ use common::sense; BEGIN { - our $VERSION = '0.91'; + our $VERSION = '0.95'; require XSLoader; XSLoader::load (__PACKAGE__, $VERSION);