[ViewVC] Diff of: cvs/OpenCL/OpenCL.pm

Comparing OpenCL/OpenCL.pm (file contents):
Revision 1.32 by root, Thu Apr 19 12:12:03 2012 UTC vs.
Revision 1.47 by root, Sat Apr 21 19:48:58 2012 UTC

…		…
160	';	160	';
161		161
162	my $prog = $ctx->program_with_source ($src);	162	my $prog = $ctx->program_with_source ($src);
163		163
164	# build croaks on compile errors, so catch it and print the compile errors	164	# build croaks on compile errors, so catch it and print the compile errors
165	eval { $prog->build ($dev); 1 }	165	eval { $prog->build ($dev, "-cl-fast-relaxed-math"); 1 }
166	or die $prog->build_log;	166	or die $prog->build_log;
167		167
168	my $kernel = $prog->kernel ("squareit");	168	my $kernel = $prog->kernel ("squareit");
169		169
170	=head2 Create some input and output float buffers, then call the	170	=head2 Create some input and output float buffers, then call the
…		…
210	# enqueue an async read	210	# enqueue an async read
211	$ev = $queue->enqueue_read_buffer ($output, 0, 0, OpenCL::SIZEOF_FLOAT * 4, my $data, $ev);	211	$ev = $queue->enqueue_read_buffer ($output, 0, 0, OpenCL::SIZEOF_FLOAT * 4, my $data, $ev);
212		212
213	# wait for the last event to complete	213	# wait for the last event to complete
214	$ev->wait;	214	$ev->wait;
		215
		216	=head2 Use the OpenGL module to share a texture between OpenCL and OpenGL and draw some julia
		217	set tunnel effect.
		218
		219	This is quite a long example to get you going.
		220
		221	use OpenGL ":all";
		222	use OpenCL;
		223
		224	# open a window and create a gl texture
		225	OpenGL::glpOpenWindow width => 256, height => 256;
		226	my $texid = glGenTextures_p 1;
		227	glBindTexture GL_TEXTURE_2D, $texid;
		228	glTexImage2D_c GL_TEXTURE_2D, 0, GL_RGBA8, 256, 256, 0, GL_RGBA, GL_UNSIGNED_BYTE, 0;
		229
		230	# find and use the first opencl device that let's us get a shared opengl context
		231	my $platform;
		232	my $dev;
		233	my $ctx;
		234
		235	for (OpenCL::platforms) {
		236	$platform = $_;
		237	for ($platform->devices) {
		238	$dev = $_;
		239	$ctx = $platform->context ([OpenCL::GLX_DISPLAY_KHR, undef, OpenCL::GL_CONTEXT_KHR, undef], [$dev])
		240	and last;
		241	}
		242	}
		243
		244	$ctx
		245	or die "cannot find suitable OpenCL device\n";
		246
		247	my $queue = $ctx->queue ($dev);
		248
		249	# now attach an opencl image2d object to the opengl texture
		250	my $tex = $ctx->gl_texture2d (OpenCL::MEM_WRITE_ONLY, GL_TEXTURE_2D, 0, $texid);
		251
		252	# now the boring opencl code
		253	my $src = <<EOF;
		254	kernel void
		255	juliatunnel (write_only image2d_t img, float time)
		256	{
		257	float2 p = (float2)(get_global_id (0), get_global_id (1)) / 256.f * 2.f - 1.f;
		258
		259	float2 m = (float2)(1.f, p.y) / fabs (p.x);
		260	m.x = fabs (fmod (m.x + time * 0.05f, 4.f)) - 2.f;
		261
		262	float2 z = m;
		263	float2 c = (float2)(sin (time * 0.05005), cos (time * 0.06001));
		264
		265	for (int i = 0; i < 25 && dot (z, z) < 4.f; ++i)
		266	z = (float2)(z.x * z.x - z.y * z.y, 2.f * z.x * z.y) + c;
		267
		268	float3 colour = (float3)(z.x, z.y, z.x * z.y);
		269	write_imagef (img, (int2)(get_global_id (0), get_global_id (1)), (float4)(colour * p.x * p.x, 1.));
		270	}
		271	EOF
		272	my $prog = $ctx->program_with_source ($src);
		273	eval { $prog->build ($dev); 1 }
		274	or die $prog->build_log ($dev);
		275
		276	my $kernel = $prog->kernel ("juliatunnel");
		277
		278	# program compiled, kernel ready, now draw and loop
		279
		280	for (my $time; ; ++$time) {
		281	# acquire objects from opengl
		282	$queue->enqueue_acquire_gl_objects ([$tex]);
		283
		284	# configure and run our kernel
		285	$kernel->set_image2d (0, $tex);
		286	$kernel->set_float (1, $time);
		287	$queue->enqueue_nd_range_kernel ($kernel, undef, [256, 256], undef);
		288
		289	# release objects to opengl again
		290	$queue->enqueue_release_gl_objects ([$tex]);
		291
		292	# wait
		293	$queue->finish;
		294
		295	# now draw the texture, the defaults should be all right
		296	glTexParameterf GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST;
		297
		298	glEnable GL_TEXTURE_2D;
		299	glBegin GL_QUADS;
		300	glTexCoord2f 0, 1; glVertex3i -1, -1, -1;
		301	glTexCoord2f 0, 0; glVertex3i 1, -1, -1;
		302	glTexCoord2f 1, 0; glVertex3i 1, 1, -1;
		303	glTexCoord2f 1, 1; glVertex3i -1, 1, -1;
		304	glEnd;
		305
		306	glXSwapBuffers;
		307
		308	select undef, undef, undef, 1/60;
		309	}
215		310
216	=head1 DOCUMENTATION	311	=head1 DOCUMENTATION
217		312
218	=head2 BASIC CONVENTIONS	313	=head2 BASIC CONVENTIONS
219		314
…		…
241	=item * Structures are often specified by flattening out their components	336	=item * Structures are often specified by flattening out their components
242	as with short vectors, and returned as arrayrefs.	337	as with short vectors, and returned as arrayrefs.
243		338
244	=item * When enqueuing commands, the wait list is specified by adding	339	=item * When enqueuing commands, the wait list is specified by adding
245	extra arguments to the function - anywhere a C<$wait_events...> argument	340	extra arguments to the function - anywhere a C<$wait_events...> argument
246	is documented this can be any number of event objects.	341	is documented this can be any number of event objects. As an extsnion
		342	implemented by this module, C<undef> values will be ignored in the event
		343	list.
247		344
248	=item * When enqueuing commands, if the enqueue method is called in void	345	=item * When enqueuing commands, if the enqueue method is called in void
249	context, no event is created. In all other contexts an event is returned	346	context, no event is created. In all other contexts an event is returned
250	by the method.	347	by the method.
251		348
…		…
271	ulong IV - Q	368	ulong IV - Q
272	float NV float f	369	float NV float f
273	half IV ushort S	370	half IV ushort S
274	double NV double d	371	double NV double d
275		372
276	=head2 OpenGL sharing	373	=head2 GLX SUPPORT
277		374
278	This module can be optionally compiled with support for	375	Due to the sad state that OpenGL support is in in Perl (mostly the OpenGL
279	OpenGL sharing. The sharing functions are only available when	376	module, which has little to no documentation and has little to no support
280	C<OpenCL::HAVE_OPENGL> returns true, otherwise they are absent and cannot	377	for glX), this module, as a special extension, treats context creation
281	be called.	378	properties C<OpenCL::GLX_DISPLAY_KHR> and C<OpenCL::GL_CONTEXT_KHR>
		379	specially: If either or both of these are C<undef>, then the OpenCL
		380	module tries to dynamically resolve C<glXGetCurrentDisplay> and
		381	C<glXGetCurrentContext>, call these functions and use their return values
		382	instead.
		383
		384	For this to work, the OpenGL library must be loaded, a GLX context must
		385	have been created and be made current, and C<dlsym> must be available and
		386	capable of finding the function via C<RTLD_DEFAULT>.
282		387
283	=head2 THE OpenCL PACKAGE	388	=head2 THE OpenCL PACKAGE
284		389
285	=over 4	390	=over 4
286		391
…		…
645		750
646	=item @device_partition_property_exts = $device->affinity_domains_ext	751	=item @device_partition_property_exts = $device->affinity_domains_ext
647		752
648	Calls C<clGetDeviceInfo> with C<CL_DEVICE_AFFINITY_DOMAINS_EXT> and returns the result.	753	Calls C<clGetDeviceInfo> with C<CL_DEVICE_AFFINITY_DOMAINS_EXT> and returns the result.
649		754
650	=item $uint = $device->reference_count_ext	755	=item $uint = $device->reference_count_ext
651		756
652	Calls C<clGetDeviceInfo> with C<CL_DEVICE_REFERENCE_COUNT_EXT > and returns the result.	757	Calls C<clGetDeviceInfo> with C<CL_DEVICE_REFERENCE_COUNT_EXT > and returns the result.
653		758
654	=item @device_partition_property_exts = $device->partition_style_ext	759	=item @device_partition_property_exts = $device->partition_style_ext
655		760
…		…
666	=item $queue = $ctx->queue ($device, $properties)	771	=item $queue = $ctx->queue ($device, $properties)
667		772
668	Create a new OpenCL::Queue object from the context and the given device.	773	Create a new OpenCL::Queue object from the context and the given device.
669		774
670	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateCommandQueue.html>	775	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateCommandQueue.html>
		776
		777	Example: create an out-of-order queue.
		778
		779	$queue = $ctx->queue ($device, OpenCL::QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE);
671		780
672	=item $ev = $ctx->user_event	781	=item $ev = $ctx->user_event
673		782
674	Creates a new OpenCL::UserEvent object.	783	Creates a new OpenCL::UserEvent object.
675		784
…		…
698		807
699	Creates a new OpenCL::Image3D object and optionally initialises it with	808	Creates a new OpenCL::Image3D object and optionally initialises it with
700	the given data values.	809	the given data values.
701		810
702	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateImage3D.html>	811	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateImage3D.html>
		812
		813	=item $buffer = $ctx->gl_buffer ($flags, $bufobj)
		814
		815	Creates a new OpenCL::Buffer (actually OpenCL::BufferObj) object that refers to the given
		816	OpenGL buffer object.
		817
		818	http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLBuffer.html
		819
		820	=item $img = $ctx->gl_texture ($flags, $target, $miplevel, $texture)
		821
		822	Creates a new OpenCL::Image object that refers to the given OpenGL
		823	texture object or buffer.
		824
		825	http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateFromGLTexture.html
		826
		827	=item $img = $ctx->gl_texture2d ($flags, $target, $miplevel, $texture)
		828
		829	Creates a new OpenCL::Image2D object that refers to the given OpenGL
		830	2D texture object.
		831
		832	http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLTexture2D.html
		833
		834	=item $img = $ctx->gl_texture3d ($flags, $target, $miplevel, $texture)
		835
		836	Creates a new OpenCL::Image3D object that refers to the given OpenGL
		837	3D texture object.
		838
		839	http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLTexture3D.html
		840
		841	=item $ctx->gl_renderbuffer ($flags, $renderbuffer)
		842
		843	Creates a new OpenCL::Image2D object that refers to the given OpenGL
		844	render buffer.
		845
		846	http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLRenderbuffer.html
703		847
704	=item @formats = $ctx->supported_image_formats ($flags, $image_type)	848	=item @formats = $ctx->supported_image_formats ($flags, $image_type)
705		849
706	Returns a list of matching image formats - each format is an arrayref with	850	Returns a list of matching image formats - each format is an arrayref with
707	two values, $channel_order and $channel_type, in it.	851	two values, $channel_order and $channel_type, in it.
…		…
758	for completion, unless the method is called in void context, in which case	902	for completion, unless the method is called in void context, in which case
759	no event object is created.	903	no event object is created.
760		904
761	They also allow you to specify any number of other event objects that this	905	They also allow you to specify any number of other event objects that this
762	request has to wait for before it starts executing, by simply passing the	906	request has to wait for before it starts executing, by simply passing the
763	event objects as extra parameters to the enqueue methods.	907	event objects as extra parameters to the enqueue methods. To simplify
		908	program design, this module ignores any C<undef> values in the list of
		909	events. This makes it possible to code operations such as this, without
		910	having to put a valid event object into C<$event> first:
		911
		912	$event = $queue->enqueue_xxx (..., $event);
764		913
765	Queues execute in-order by default, without any parallelism, so in most	914	Queues execute in-order by default, without any parallelism, so in most
766	cases (i.e. you use only one queue) it's not necessary to wait for or	915	cases (i.e. you use only one queue) it's not necessary to wait for or
767	create event objects.	916	create event objects, althoguh an our of order queue is often a bit
		917	faster.
768		918
769	=over 4	919	=over 4
770		920
771	=item $ev = $queue->enqueue_read_buffer ($buffer, $blocking, $offset, $len, $data, $wait_events...)	921	=item $ev = $queue->enqueue_read_buffer ($buffer, $blocking, $offset, $len, $data, $wait_events...)
772		922
…		…
838	reference to an array of local work sizes, with the same number of	988	reference to an array of local work sizes, with the same number of
839	elements as @$global_work_size.	989	elements as @$global_work_size.
840		990
841	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueNDRangeKernel.html>	991	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueNDRangeKernel.html>
842		992
843	=item $ev = $queue->enqueue_marker	993	=item $ev = $queue->enqueue_acquire_gl_objects ([object, ...], $wait_events...)
844		994
		995	Enqueues a list (an array-ref of OpenCL::Memory objects) to be acquired
		996	for subsequent OpenCL usage.
		997
845	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueMarker.html>	998	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueAcquireGLObjects.html>
		999
		1000	=item $ev = $queue->enqueue_release_gl_objects ([object, ...], $wait_events...)
		1001
		1002	Enqueues a list (an array-ref of OpenCL::Memory objects) to be released
		1003	for subsequent OpenGL usage.
		1004
		1005	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReleaseGLObjects.html>
846		1006
847	=item $ev = $queue->enqueue_wait_for_events ($wait_events...)	1007	=item $ev = $queue->enqueue_wait_for_events ($wait_events...)
848		1008
849	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWaitForEvents.html>	1009	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWaitForEvents.html>
850		1010
851	=item $queue->enqueue_barrier	1011	=item $ev = $queue->enqueue_marker ($wait_events...)
852		1012
		1013	L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueMarkerWithWaitList.html>
		1014
		1015	=item $ev = $queue->enqueue_barrier ($wait_events...)
		1016
853	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueBarrier.html>	1017	L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueBarrierWithWaitList.html>
854		1018
855	=item $queue->flush	1019	=item $queue->flush
856		1020
857	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFlush.html>	1021	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFlush.html>
858		1022
…		…
938	=item $int = $mem->offset	1102	=item $int = $mem->offset
939		1103
940	Calls C<clGetMemObjectInfo> with C<CL_MEM_OFFSET> and returns the result.	1104	Calls C<clGetMemObjectInfo> with C<CL_MEM_OFFSET> and returns the result.
941		1105
942	=for gengetinfo end mem	1106	=for gengetinfo end mem
		1107
		1108	=item ($type, $name) = $mem->gl_object_info
		1109
		1110	Returns the OpenGL object type (e.g. OpenCL::GL_OBJECT_TEXTURE2D) and the
		1111	object "name" (e.g. the texture name) used to create this memory object.
		1112
		1113	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetGLObjectInfo.html>
943		1114
944	=back	1115	=back
945		1116
946	=head2 THE OpenCL::Buffer CLASS	1117	=head2 THE OpenCL::Buffer CLASS
947		1118
…		…
966		1137
967	=back	1138	=back
968		1139
969	=head2 THE OpenCL::Image CLASS	1140	=head2 THE OpenCL::Image CLASS
970		1141
971	This is the superclass of all image objects - OpenCL::Image2D and OpenCL::Image3D.	1142	This is the superclass of all image objects - OpenCL::Image1D,
		1143	OpenCL::Image1DArray, OpenCL::Image1DBuffer, OpenCL::Image2D,
		1144	OpenCL::Image2DArray and OpenCL::Image3D.
972		1145
973	=over 4	1146	=over 4
974		1147
975	=item $packed_value = $ev->image_info ($name)	1148	=item $packed_value = $ev->image_info ($name)
976		1149
…		…
1006	=item $int = $image->depth	1179	=item $int = $image->depth
1007		1180
1008	Calls C<clGetImageInfo> with C<CL_IMAGE_DEPTH> and returns the result.	1181	Calls C<clGetImageInfo> with C<CL_IMAGE_DEPTH> and returns the result.
1009		1182
1010	=for gengetinfo end image	1183	=for gengetinfo end image
		1184
		1185	=for gengetinfo begin gl_texture
		1186
		1187	=item $GLenum = $gl_texture->target
		1188
		1189	Calls C<clGetGLTextureInfo> with C<CL_GL_TEXTURE_TARGET> and returns the result.
		1190
		1191	=item $GLint = $gl_texture->gl_mipmap_level
		1192
		1193	Calls C<clGetGLTextureInfo> with C<CL_GL_MIPMAP_LEVEL> and returns the result.
		1194
		1195	=for gengetinfo end gl_texture
1011		1196
1012	=back	1197	=back
1013		1198
1014	=head2 THE OpenCL::Sampler CLASS	1199	=head2 THE OpenCL::Sampler CLASS
1015		1200
…		…
1205		1390
1206	This is a family of methods to set the kernel argument with the number C<$index> to the give C<$value>.	1391	This is a family of methods to set the kernel argument with the number C<$index> to the give C<$value>.
1207		1392
1208	TYPE is one of C<char>, C<uchar>, C<short>, C<ushort>, C<int>, C<uint>,	1393	TYPE is one of C<char>, C<uchar>, C<short>, C<ushort>, C<int>, C<uint>,
1209	C<long>, C<ulong>, C<half>, C<float>, C<double>, C<memory>, C<buffer>,	1394	C<long>, C<ulong>, C<half>, C<float>, C<double>, C<memory>, C<buffer>,
1210	C<image2d>, C<image3d>, C<sampler> or C<event>.	1395	C<image2d>, C<image3d>, C<sampler>, C<local> or C<event>.
1211		1396
1212	Chars and integers (including the half type) are specified as integers,	1397	Chars and integers (including the half type) are specified as integers,
1213	float and double as floating point values, memory/buffer/image2d/image3d	1398	float and double as floating point values, memory/buffer/image2d/image3d
1214	must be an object of that type or C<undef>, and sampler and event must be	1399	must be an object of that type or C<undef>, local-memory arguments are
1215	objects of that type.	1400	set by specifying the size, and sampler and event must be objects of that
		1401	type.
1216		1402
1217	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetKernelArg.html>	1403	L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetKernelArg.html>
1218		1404
1219	=back	1405	=back
1220		1406
…		…
1309	package OpenCL;	1495	package OpenCL;
1310		1496
1311	use common::sense;	1497	use common::sense;
1312		1498
1313	BEGIN {	1499	BEGIN {
1314	our $VERSION = '0.92';	1500	our $VERSION = '0.96';
1315		1501
1316	require XSLoader;	1502	require XSLoader;
1317	XSLoader::load (__PACKAGE__, $VERSION);	1503	XSLoader::load (__PACKAGE__, $VERSION);
1318		1504
1319	@OpenCL::Buffer::ISA =	1505	@OpenCL::Buffer::ISA =
1320	@OpenCL::Image::ISA = OpenCL::Memory::;	1506	@OpenCL::Image::ISA = OpenCL::Memory::;
1321		1507
1322	@OpenCL::BufferObj::ISA = OpenCL::Buffer::;	1508	@OpenCL::BufferObj::ISA = OpenCL::Buffer::;
1323		1509
1324	@OpenCL::Image2D::ISA =	1510	@OpenCL::Image2D::ISA =
		1511	@OpenCL::Image3D::ISA =
		1512	@OpenCL::Image2DArray::ISA =
		1513	@OpenCL::Image1D::ISA =
		1514	@OpenCL::Image1DArray::ISA =
1325	@OpenCL::Image3D::ISA = OpenCL::Image::;	1515	@OpenCL::Image1DBuffer::ISA = OpenCL::Image::;
1326		1516
1327	@OpenCL::UserEvent::ISA = OpenCL::Event::;	1517	@OpenCL::UserEvent::ISA = OpenCL::Event::;
1328	}	1518	}
1329		1519
1330	1;	1520	1;
1331		1521
1332	=head1 AUTHOR	1522	=head1 AUTHOR

Diff Legend

-–
+Removed lines
-+
+Added lines
-<
+Changed lines
->
+Changed lines

Comparing OpenCL/OpenCL.pm (file contents): Revision 1.32 by root, Thu Apr 19 12:12:03 2012 UTC vs. Revision 1.47 by root, Sat Apr 21 19:48:58 2012 UTC

Diff Legend

Comparing OpenCL/OpenCL.pm (file contents):
Revision 1.32 by root, Thu Apr 19 12:12:03 2012 UTC vs.
Revision 1.47 by root, Sat Apr 21 19:48:58 2012 UTC