ViewVC Help
View File | Revision Log | Show Annotations | Download File
/cvs/OpenCL/OpenCL.xs
(Generate patch)

Comparing OpenCL/OpenCL.xs (file contents):
Revision 1.3 by root, Tue Nov 15 20:38:07 2011 UTC vs.
Revision 1.7 by root, Thu Nov 17 02:10:39 2011 UTC

… …
20typedef cl_mem OpenCL__Image3D_ornull; 20typedef cl_mem OpenCL__Image3D_ornull;
21typedef cl_sampler OpenCL__Sampler; 21typedef cl_sampler OpenCL__Sampler;
22typedef cl_program OpenCL__Program; 22typedef cl_program OpenCL__Program;
23typedef cl_kernel OpenCL__Kernel; 23typedef cl_kernel OpenCL__Kernel;
24typedef cl_event OpenCL__Event; 24typedef cl_event OpenCL__Event;
25typedef cl_event OpenCL__UserEvent;
26
27typedef SV *FUTURE;
28
29/*****************************************************************************/
30
31/* up to two temporary buffers */
32static void *
33tmpbuf (size_t size)
34{
35 static int idx;
36 static void *buf [2];
37 static size_t len [2];
38
39 idx ^= 1;
40
41 if (len [idx] < size)
42 {
43 free (buf [idx]);
44 len [idx] = ((size + 31) & ~4095) + 4096 - 32;
45 buf [idx] = malloc (len [idx]);
46 }
47
48 return buf [idx];
49}
50
51/*****************************************************************************/
25 52
26typedef struct 53typedef struct
27{ 54{
28 IV iv; 55 IV iv;
29 const char *name; 56 const char *name;
… …
63 }; 90 };
64 91
65 return iv2str (err, errstr, sizeof (errstr) / sizeof (errstr [0]), "ERROR(%d)"); 92 return iv2str (err, errstr, sizeof (errstr) / sizeof (errstr [0]), "ERROR(%d)");
66} 93}
67 94
95/*****************************************************************************/
96
97static cl_int last_error;
98
68#define FAIL(name,res) \ 99#define FAIL(name,err) \
69 croak ("cl" # name ": %s", err2str (res)); 100 croak ("cl" # name ": %s", err2str (last_error = err));
70 101
71#define NEED_SUCCESS(name,args) \ 102#define NEED_SUCCESS(name,args) \
72 do { \ 103 do { \
73 cl_int res = cl ## name args; \ 104 cl_int res = cl ## name args; \
74 \ 105 \
75 if (res) \ 106 if (res) \
76 FAIL (name, res); \ 107 FAIL (name, res); \
77 } while (0) 108 } while (0)
78 109
110/*****************************************************************************/
111
79#define NEW_MORTAL_OBJ(class,ptr) sv_setref_pv (sv_newmortal (), class, ptr) 112#define NEW_MORTAL_OBJ(class,ptr) sv_setref_pv (sv_newmortal (), class, ptr)
80#define XPUSH_NEW_OBJ(class,ptr) XPUSHs (NEW_MORTAL_OBJ (class, ptr)) 113#define XPUSH_NEW_OBJ(class,ptr) XPUSHs (NEW_MORTAL_OBJ (class, ptr))
81 114
82/*TODO*/ 115static void *
83#define EVENT_LIST(items,count) cl_uint event_list_count = 0; cl_event *event_list_ptr = 0 116SvPTROBJ (const char *func, const char *svname, SV *sv, const char *pkg)
117{
118 if (SvROK (sv) && sv_derived_from (sv, pkg))
119 return (void *)SvIV (SvRV (sv));
120
121 croak ("%s: %s is not of type %s", func, svname, pkg);
122}
123
124/*****************************************************************************/
125
126static cl_event *
127event_list (SV **items, int count)
128{
129 cl_event *list = tmpbuf (sizeof (cl_event) * count);
130
131 while (count--)
132 list [count] = SvPTROBJ ("clEnqueue", "wait_events", items [count], "OpenCL::Event");
133
134 return list;
135}
136
137#define EVENT_LIST(items,count) \
138 cl_uint event_list_count = (count); \
139 cl_event *event_list_ptr = event_list (&ST (items), event_list_count)
84 140
85#define INFO(class) \ 141#define INFO(class) \
86{ \ 142{ \
87 size_t size; \ 143 size_t size; \
88 SV *sv; \ 144 SV *sv; \
… …
94 SvCUR_set (sv, size); \ 150 SvCUR_set (sv, size); \
95 NEED_SUCCESS (Get ## class ## Info, (this, name, size, SvPVX (sv), 0)); \ 151 NEED_SUCCESS (Get ## class ## Info, (this, name, size, SvPVX (sv), 0)); \
96 XPUSHs (sv); \ 152 XPUSHs (sv); \
97} 153}
98 154
99static void *
100SvPTROBJ (const char *func, const char *svname, SV *sv, const char *pkg)
101{
102 if (SvROK (sv) && sv_derived_from (sv, pkg))
103 return (void *)SvIV (SvRV (sv));
104
105 croak ("%s: %s is not of type %s", func, svname, pkg);
106}
107
108MODULE = OpenCL PACKAGE = OpenCL 155MODULE = OpenCL PACKAGE = OpenCL
109 156
110PROTOTYPES: ENABLE 157PROTOTYPES: ENABLE
111 158
112BOOT: 159BOOT:
113{ 160{
114 HV *stash = gv_stashpv ("OpenCL", 1); 161 HV *stash = gv_stashpv ("OpenCL", 1);
115 static const ivstr *civ, const_iv[] = { 162 static const ivstr *civ, const_iv[] = {
116 { sizeof (cl_char ), "SIZEOF_CHAR" }, 163 { sizeof (cl_char ), "SIZEOF_CHAR" },
117 { sizeof (cl_uchar ), "SIZEOF_UCHAR" }, 164 { sizeof (cl_uchar ), "SIZEOF_UCHAR" },
118 { sizeof (cl_short ), "SIZEOF_SHORT" }, 165 { sizeof (cl_short ), "SIZEOF_SHORT" },
119 { sizeof (cl_ushort), "SIZEOF_USHORT"}, 166 { sizeof (cl_ushort), "SIZEOF_USHORT" },
120 { sizeof (cl_int ), "SIZEOF_INT" }, 167 { sizeof (cl_int ), "SIZEOF_INT" },
121 { sizeof (cl_uint ), "SIZEOF_UINT" }, 168 { sizeof (cl_uint ), "SIZEOF_UINT" },
122 { sizeof (cl_long ), "SIZEOF_LONG" }, 169 { sizeof (cl_long ), "SIZEOF_LONG" },
123 { sizeof (cl_ulong ), "SIZEOF_ULONG" }, 170 { sizeof (cl_ulong ), "SIZEOF_ULONG" },
124 { sizeof (cl_half ), "SIZEOF_HALF" }, 171 { sizeof (cl_half ), "SIZEOF_HALF" },
125 { sizeof (cl_float ), "SIZEOF_FLOAT" }, 172 { sizeof (cl_float ), "SIZEOF_FLOAT" },
173 { sizeof (cl_double), "SIZEOF_DOUBLE" },
126#include "constiv.h" 174#include "constiv.h"
127 }; 175 };
128 for (civ = const_iv + sizeof (const_iv) / sizeof (const_iv [0]); civ > const_iv; civ--) 176 for (civ = const_iv + sizeof (const_iv) / sizeof (const_iv [0]); civ > const_iv; civ--)
129 newCONSTSUB (stash, (char *)civ[-1].name, newSViv (civ[-1].iv)); 177 newCONSTSUB (stash, (char *)civ[-1].name, newSViv (civ[-1].iv));
130} 178}
131 179
180cl_int
181errno ()
182 CODE:
183 errno = last_error;
184
132const char * 185const char *
133err2str (cl_int err) 186err2str (cl_int err)
134 187
135const char * 188const char *
136enum2str (cl_uint value) 189enum2str (cl_uint value)
… …
142 cl_platform_id *list; 195 cl_platform_id *list;
143 cl_uint count; 196 cl_uint count;
144 int i; 197 int i;
145 198
146 NEED_SUCCESS (GetPlatformIDs, (0, 0, &count)); 199 NEED_SUCCESS (GetPlatformIDs, (0, 0, &count));
147 Newx (list, count, cl_platform_id); 200 list = tmpbuf (sizeof (*list) * count);
148 NEED_SUCCESS (GetPlatformIDs, (count, list, 0)); 201 NEED_SUCCESS (GetPlatformIDs, (count, list, 0));
149 202
150 EXTEND (SP, count); 203 EXTEND (SP, count);
151 for (i = 0; i < count; ++i) 204 for (i = 0; i < count; ++i)
152 PUSHs (NEW_MORTAL_OBJ ("OpenCL::Platform", list [i])); 205 PUSHs (NEW_MORTAL_OBJ ("OpenCL::Platform", list [i]));
153
154 Safefree (list);
155} 206}
156 207
157void 208void
158context_from_type_simple (cl_device_type type = CL_DEVICE_TYPE_DEFAULT) 209context_from_type (FUTURE properties = 0, cl_device_type type = CL_DEVICE_TYPE_DEFAULT, FUTURE notify = 0)
159 PPCODE: 210 PPCODE:
160{ 211{
161 cl_int res; 212 cl_int res;
162 cl_context ctx = clCreateContextFromType (0, type, 0, 0, &res); 213 cl_context ctx = clCreateContextFromType (0, type, 0, 0, &res);
163 214
… …
191 cl_device_id *list; 242 cl_device_id *list;
192 cl_uint count; 243 cl_uint count;
193 int i; 244 int i;
194 245
195 NEED_SUCCESS (GetDeviceIDs, (this, type, 0, 0, &count)); 246 NEED_SUCCESS (GetDeviceIDs, (this, type, 0, 0, &count));
196 Newx (list, count, cl_device_id); 247 list = tmpbuf (sizeof (*list) * count);
197 NEED_SUCCESS (GetDeviceIDs, (this, type, count, list, 0)); 248 NEED_SUCCESS (GetDeviceIDs, (this, type, count, list, 0));
198 249
199 EXTEND (SP, count); 250 EXTEND (SP, count);
200 for (i = 0; i < count; ++i) 251 for (i = 0; i < count; ++i)
201 PUSHs (sv_setref_pv (sv_newmortal (), "OpenCL::Device", list [i])); 252 PUSHs (sv_setref_pv (sv_newmortal (), "OpenCL::Device", list [i]));
202
203 Safefree (list);
204} 253}
205 254
206void 255void
207context_from_type_simple (OpenCL::Platform this, cl_device_type type = CL_DEVICE_TYPE_DEFAULT) 256context_from_type (OpenCL::Platform this, FUTURE properties = 0, cl_device_type type = CL_DEVICE_TYPE_DEFAULT, FUTURE notify = 0)
208 PPCODE: 257 PPCODE:
209{ 258{
210 cl_int res; 259 cl_int res;
211 cl_context_properties props[] = { CL_CONTEXT_PLATFORM, (cl_context_properties)this, 0 }; 260 cl_context_properties props[] = { CL_CONTEXT_PLATFORM, (cl_context_properties)this, 0 };
212 cl_context ctx = clCreateContextFromType (props, type, 0, 0, &res); 261 cl_context ctx = clCreateContextFromType (props, type, 0, 0, &res);
… …
223info (OpenCL::Device this, cl_device_info name) 272info (OpenCL::Device this, cl_device_info name)
224 PPCODE: 273 PPCODE:
225 INFO (Device) 274 INFO (Device)
226 275
227void 276void
228context_simple (OpenCL::Device this) 277context (OpenCL::Device this, FUTURE properties = 0, FUTURE notify = 0)
229 PPCODE: 278 PPCODE:
230{ 279{
231 cl_int res; 280 cl_int res;
232 cl_context ctx = clCreateContext (0, 1, &this, 0, 0, &res); 281 cl_context ctx = clCreateContext (0, 1, &this, 0, 0, &res);
233 282
… …
248info (OpenCL::Context this, cl_context_info name) 297info (OpenCL::Context this, cl_context_info name)
249 PPCODE: 298 PPCODE:
250 INFO (Context) 299 INFO (Context)
251 300
252void 301void
253command_queue_simple (OpenCL::Context this, OpenCL::Device device) 302queue (OpenCL::Context this, OpenCL::Device device, cl_command_queue_properties properties = 0)
254 PPCODE: 303 PPCODE:
255{ 304{
256 cl_int res; 305 cl_int res;
257 cl_command_queue queue = clCreateCommandQueue (this, device, 0, &res); 306 cl_command_queue queue = clCreateCommandQueue (this, device, properties, &res);
258 307
259 if (res) 308 if (res)
260 FAIL (CreateCommandQueue, res); 309 FAIL (CreateCommandQueue, res);
261 310
262 XPUSH_NEW_OBJ ("OpenCL::Queue", queue); 311 XPUSH_NEW_OBJ ("OpenCL::Queue", queue);
312}
313
314void
315user_event (OpenCL::Context this)
316 PPCODE:
317{
318 cl_int res;
319 cl_event ev = clCreateUserEvent (this, &res);
320
321 if (res)
322 FAIL (CreateUserevent, res);
323
324 XPUSH_NEW_OBJ ("OpenCL::UserEvent", ev);
263} 325}
264 326
265void 327void
266buffer (OpenCL::Context this, cl_mem_flags flags, size_t len) 328buffer (OpenCL::Context this, cl_mem_flags flags, size_t len)
267 PPCODE: 329 PPCODE:
… …
558 620
559 if (ev) 621 if (ev)
560 XPUSH_NEW_OBJ ("OpenCL::Event", ev); 622 XPUSH_NEW_OBJ ("OpenCL::Event", ev);
561} 623}
562 624
563 /*TODO http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueNDRangeKernel.html */ 625void
626enqueue_nd_range_kernel (OpenCL::Queue this, OpenCL::Kernel kernel, SV *global_work_offset, SV *global_work_size, SV *local_work_size = &PL_sv_undef, ...)
627 PPCODE:
628{
629 cl_event ev = 0;
630 size_t *gwo = 0, *gws, *lws = 0;
631 int gws_len;
632 size_t *lists;
633 int i;
634 EVENT_LIST (5, items - 5);
635
636 if (!SvROK (global_work_size) || SvTYPE (SvRV (global_work_size)) != SVt_PVAV)
637 croak ("clEnqueueNDRangeKernel: global_work_size must be an array reference");
638
639 gws_len = AvFILLp (SvRV (global_work_size)) + 1;
640
641 lists = tmpbuf (sizeof (size_t) * 3 * gws_len);
642
643 gws = lists + gws_len * 0;
644 for (i = 0; i < gws_len; ++i)
645 gws [i] = SvIV (AvARRAY (SvRV (global_work_size))[i]);
646
647 if (SvOK (global_work_offset))
648 {
649 if (!SvROK (global_work_offset) || SvTYPE (SvRV (global_work_offset)) != SVt_PVAV)
650 croak ("clEnqueueNDRangeKernel: global_work_offset must be undef or an array reference");
651
652 if (AvFILLp (SvRV (global_work_size)) + 1 != gws_len)
653 croak ("clEnqueueNDRangeKernel: global_work_offset must be undef or an array of same size as global_work_size");
654
655 gwo = lists + gws_len * 1;
656 for (i = 0; i < gws_len; ++i)
657 gwo [i] = SvIV (AvARRAY (SvRV (global_work_offset))[i]);
658 }
659
660 if (SvOK (local_work_size))
661 {
662 if (SvOK (local_work_size) && !SvROK (local_work_size) || SvTYPE (SvRV (local_work_size)) != SVt_PVAV)
663 croak ("clEnqueueNDRangeKernel: global_work_size must be undef or an array reference");
664
665 if (AvFILLp (SvRV (local_work_size)) + 1 != gws_len)
666 croak ("clEnqueueNDRangeKernel: local_work_local must be undef or an array of same size as global_work_size");
667
668 lws = lists + gws_len * 2;
669 for (i = 0; i < gws_len; ++i)
670 lws [i] = SvIV (AvARRAY (SvRV (local_work_size))[i]);
671 }
672
673 NEED_SUCCESS (EnqueueNDRangeKernel, (this, kernel, gws_len, gwo, gws, lws, event_list_count, event_list_ptr, GIMME_V != G_VOID ? &ev : 0));
674
675 if (ev)
676 XPUSH_NEW_OBJ ("OpenCL::Event", ev);
677}
564 678
565void 679void
566enqueue_marker (OpenCL::Queue this) 680enqueue_marker (OpenCL::Queue this)
567 PPCODE: 681 PPCODE:
568{ 682{
… …
725set_float (OpenCL::Kernel this, cl_uint idx, cl_float value) 839set_float (OpenCL::Kernel this, cl_uint idx, cl_float value)
726 CODE: 840 CODE:
727 clSetKernelArg (this, idx, sizeof (value), &value); 841 clSetKernelArg (this, idx, sizeof (value), &value);
728 842
729void 843void
844set_double (OpenCL::Kernel this, cl_uint idx, cl_double value)
845 CODE:
846 clSetKernelArg (this, idx, sizeof (value), &value);
847
848void
730set_memory (OpenCL::Kernel this, cl_uint idx, OpenCL::Memory_ornull value) 849set_memory (OpenCL::Kernel this, cl_uint idx, OpenCL::Memory_ornull value)
731 CODE: 850 CODE:
732 clSetKernelArg (this, idx, sizeof (value), &value); 851 clSetKernelArg (this, idx, sizeof (value), &value);
733 852
734void 853void
… …
771void 890void
772wait (OpenCL::Event this) 891wait (OpenCL::Event this)
773 CODE: 892 CODE:
774 clWaitForEvents (1, &this); 893 clWaitForEvents (1, &this);
775 894
895MODULE = OpenCL PACKAGE = OpenCL::UserEvent
896
897void
898set_status (OpenCL::UserEvent this, cl_int execution_status)
899 CODE:
900 clSetUserEventStatus (this, execution_status);
901

Diff Legend

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