resolve conflicts

34 files changed, 1501 insertions, 377 deletions

diff --git a/resources/cl-common.cfg b/resources/cl-common.cfg
index 53537c63..baac28f8 100644
--- a/resources/cl-common.cfg
+++ b/resources/cl-common.cfg
@@ -28,8 +28,8 @@ Ignore CL_LONG_MIN
 Ignore CL_ULONG_MAX
 
 
-#use CLException instead of RunTimeException
-RuntimeExceptionType CLException
+#use IllegalArgumentException for argument validation
+RuntimeExceptionType IllegalArgumentException
 
 #enforce client side "good behavior" by generating direct-memory-only bindings for
 #performance critical functions.
diff --git a/resources/cl-if.cfg b/resources/cl-if.cfg
index f65a417b..aa7de799 100644
--- a/resources/cl-if.cfg
+++ b/resources/cl-if.cfg
@@ -23,17 +23,17 @@ Ignore CL_GL_.*|cl.*GL.*
 Ignore clCreateContext
 CustomJavaCode CL
 CustomJavaCode CL  /** Interface to C language function: <br> <code> cl_context {@native clCreateContext}(intptr_t * , uint32_t, cl_device_id * , void (*pfn_notify)(const char *, const void *, size_t, void *), void *, int32_t * ); </code>    */
-CustomJavaCode CL  public long clCreateContext(PointerBuffer properties, PointerBuffer devices, CLErrorHandler pfn_notify, IntBuffer errcode_ret);
+CustomJavaCode CL  public long clCreateContext(NativeSizeBuffer properties, NativeSizeBuffer devices, CLErrorHandler pfn_notify, IntBuffer errcode_ret);
 
 Ignore clCreateContextFromType
 CustomJavaCode CL
 CustomJavaCode CL  /** Interface to C language function: <br> <code> cl_context {@native clCreateContextFromType}(cl_context_properties *properties, cl_device_type device_type, void (*pfn_notify)(const char *errinfo,  const void *private_info, size_t cb, void *user_data), void *user_data, cl_int *errcode_ret) ; </code>    */
-CustomJavaCode CL  public long clCreateContextFromType(PointerBuffer properties, long device_type, CLErrorHandler pfn_notify, IntBuffer errcode_ret);
+CustomJavaCode CL  public long clCreateContextFromType(NativeSizeBuffer properties, long device_type, CLErrorHandler pfn_notify, IntBuffer errcode_ret);
 
 Ignore clBuildProgram
 CustomJavaCode CL
 CustomJavaCode CL  /** Interface to C language function: <br> <code> int32_t {@native clBuildProgram}(cl_program, uint32_t, cl_device_id * , const char * , void (*pfn_notify)(cl_program, void *user_data), void * ); </code>    */
-CustomJavaCode CL  public int clBuildProgram(long program, int deviceCount, PointerBuffer devices, String options, BuildProgramCallback cb);
+CustomJavaCode CL  public int clBuildProgram(long program, int deviceCount, NativeSizeBuffer devices, String options, BuildProgramCallback cb);
 
 Ignore clSetEventCallback
 CustomJavaCode CL
diff --git a/resources/clImplCustomCode.c b/resources/clImplCustomCode.c
index f0ff3651..789491c6 100644
--- a/resources/clImplCustomCode.c
+++ b/resources/clImplCustomCode.c
@@ -299,7 +299,7 @@ Java_com_jogamp_opencl_impl_CLImpl_clBuildProgram0(JNIEnv *env, jobject _unused,
 
 /*   Java->C glue code:
  *   Java package: com.jogamp.opencl.impl.CLImpl
- *    Java method: java.nio.ByteBuffer clEnqueueMapImage(long command_queue, long image, int blocking_map, long map_flags, com.jogamp.gluegen.runtime.PointerBuffer origin, com.jogamp.gluegen.runtime.PointerBuffer range, com.jogamp.gluegen.runtime.PointerBuffer image_row_pitch, com.jogamp.gluegen.runtime.PointerBuffer image_slice_pitch, int num_events_in_wait_list, com.jogamp.gluegen.runtime.PointerBuffer event_wait_list, com.jogamp.gluegen.runtime.PointerBuffer event, java.nio.IntBuffer errcode_ret)
+ *    Java method: java.nio.ByteBuffer clEnqueueMapImage(long command_queue, long image, int blocking_map, long map_flags, com.jogamp.gluegen.common.nio.NativeSizeBuffer origin, com.jogamp.gluegen.common.nio.NativeSizeBuffer range, com.jogamp.gluegen.common.nio.NativeSizeBuffer image_row_pitch, com.jogamp.gluegen.common.nio.NativeSizeBuffer image_slice_pitch, int num_events_in_wait_list, com.jogamp.gluegen.common.nio.NativeSizeBuffer event_wait_list, com.jogamp.gluegen.common.nio.NativeSizeBuffer event, java.nio.IntBuffer errcode_ret)
  *     C function: void *  clEnqueueMapImage(cl_command_queue command_queue, cl_mem image, uint32_t blocking_map, uint64_t map_flags, const size_t * , const size_t * , size_t *  image_row_pitch, size_t *  image_slice_pitch, uint32_t num_events_in_wait_list, cl_event *  event_wait_list, cl_event *  event, int32_t *  errcode_ret);
  */
 JNIEXPORT jobject JNICALL
diff --git a/src/com/jogamp/opencl/CLBuffer.java b/src/com/jogamp/opencl/CLBuffer.java
index 57fba461..6f00247c 100644
--- a/src/com/jogamp/opencl/CLBuffer.java
+++ b/src/com/jogamp/opencl/CLBuffer.java
@@ -30,7 +30,7 @@ package com.jogamp.opencl;
 
 import com.jogamp.common.nio.Buffers;
 import java.util.List;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import com.jogamp.opencl.CLMemory.Mem;
 import java.nio.Buffer;
 import java.util.ArrayList;
@@ -108,7 +108,7 @@ public class CLBuffer<B extends Buffer> extends CLMemory<B> {
             size *= elemSize;
         }
 
-        PointerBuffer info = PointerBuffer.allocateDirect(2);
+        NativeSizeBuffer info = NativeSizeBuffer.allocateDirect(2);
         info.put(offset).put(size).rewind();
         int bitset = Mem.flagsToInt(flags);
         
diff --git a/src/com/jogamp/opencl/CLCommandQueue.java b/src/com/jogamp/opencl/CLCommandQueue.java
index ae279fa6..cbdd59fa 100644
--- a/src/com/jogamp/opencl/CLCommandQueue.java
+++ b/src/com/jogamp/opencl/CLCommandQueue.java
@@ -30,7 +30,8 @@ package com.jogamp.opencl;
 
 import com.jogamp.common.nio.CachedBufferFactory;
 import com.jogamp.opencl.gl.CLGLI;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
+import com.jogamp.opencl.gl.CLGLObject;
 import java.nio.Buffer;
 import java.nio.ByteBuffer;
 import java.nio.IntBuffer;
@@ -65,9 +66,9 @@ public class CLCommandQueue extends CLObject implements CLResource {
      * Those direct memory buffers are used to move data between the JVM and OpenCL.
      */
     private final IntBuffer pbA;
-    private final PointerBuffer ibA;
-    private final PointerBuffer ibB;
-    private final PointerBuffer ibC;
+    private final NativeSizeBuffer ibA;
+    private final NativeSizeBuffer ibB;
+    private final NativeSizeBuffer ibC;
 
     private CLCommandQueue(CLContext context, long id, CLDevice device, long properties) {
         super(context, id);
@@ -78,9 +79,9 @@ public class CLCommandQueue extends CLObject implements CLResource {
         int pbsize = PointerBuffer.ELEMENT_SIZE;
         CachedBufferFactory factory = CachedBufferFactory.create(9*pbsize + 4, true);
         
-        this.ibA = PointerBuffer.wrap(factory.newDirectByteBuffer(3*pbsize));
-        this.ibB = PointerBuffer.wrap(factory.newDirectByteBuffer(3*pbsize));
-        this.ibC = PointerBuffer.wrap(factory.newDirectByteBuffer(3*pbsize));
+        this.ibA = NativeSizeBuffer.wrap(factory.newDirectByteBuffer(3*pbsize));
+        this.ibB = NativeSizeBuffer.wrap(factory.newDirectByteBuffer(3*pbsize));
+        this.ibC = NativeSizeBuffer.wrap(factory.newDirectByteBuffer(3*pbsize));
         
         this.pbA = factory.newDirectIntBuffer(1);
 
@@ -116,7 +117,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
      */
     public CLCommandQueue putWriteBuffer(CLBuffer<?> writeBuffer, boolean blockingWrite, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -160,7 +161,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
      */
     public CLCommandQueue putReadBuffer(CLBuffer<?> readBuffer, boolean blockingRead, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -209,7 +210,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
      */
     public CLCommandQueue putCopyBuffer(CLBuffer<?> src, CLBuffer<?> dest, int srcOffset, int destOffset, long bytesToCopy, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -279,7 +280,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
             long rowPitch, long slicePitch, long hostRowPitch, long hostSlicePitch,
             boolean blockingWrite, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -357,7 +358,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
             long rowPitch, long slicePitch, long hostRowPitch, long hostSlicePitch,
             boolean blockingRead, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -438,7 +439,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
             long srcRowPitch, long srcSlicePitch, long destRowPitch, long destSlicePitch,
             CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -506,7 +507,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
     public CLCommandQueue putWriteImage(CLImage2d<?> writeImage, int inputRowPitch,
             int originX, int originY, int rangeX, int rangeY, boolean blockingWrite, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -568,7 +569,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
     public CLCommandQueue putWriteImage(CLImage3d<?> writeImage, int inputRowPitch, int inputSlicePitch,
             int originX, int originY, int originZ, int rangeX, int rangeY, int rangeZ, boolean blockingWrite, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -629,7 +630,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
     public CLCommandQueue putReadImage(CLImage2d<?> readImage, int inputRowPitch,
             int originX, int originY, int rangeX, int rangeY, boolean blockingRead, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -691,7 +692,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
     public CLCommandQueue putReadImage(CLImage3d<?> readImage, int inputRowPitch, int inputSlicePitch,
             int originX, int originY, int originZ, int rangeX, int rangeY, int rangeZ, boolean blockingRead, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -755,7 +756,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
                                         int dstOriginX, int dstOriginY,
                                         int rangeX, int rangeY, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -824,7 +825,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
                                         int dstOriginX, int dstOriginY, int dstOriginZ,
                                         int rangeX, int rangeY, int rangeZ, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -888,7 +889,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
                                         long srcOffset, int dstOriginX, int dstOriginY,
                                         int rangeX, int rangeY, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -955,7 +956,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
                                         long srcOffset, int dstOriginX, int dstOriginY, int dstOriginZ,
                                         int rangeX, int rangeY, int rangeZ, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1019,7 +1020,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
                                         int srcOriginX, int srcOriginY,
                                         int rangeX, int rangeY, long dstOffset, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1086,7 +1087,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
                                         int srcOriginX, int srcOriginY, int srcOriginZ, 
                                         int rangeX, int rangeY, int rangeZ, long dstOffset, CLEventList condition, CLEventList events) {
         
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1144,7 +1145,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
      */
     public ByteBuffer putMapBuffer(CLBuffer<?> buffer, CLMemory.Map flag, long offset, long length, boolean blockingMap, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1204,7 +1205,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
                                     int offsetX, int offsetY,
                                     int rangeX, int rangeY, boolean blockingMap, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1270,7 +1271,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
                                     int offsetX, int offsetY, int offsetZ,
                                     int rangeX, int rangeY, int rangeZ, boolean blockingMap, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1314,7 +1315,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
      */
     public CLCommandQueue putUnmapMemory(CLMemory<?> memory, Buffer mapped, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1351,7 +1352,6 @@ public class CLCommandQueue extends CLObject implements CLResource {
     public CLCommandQueue putWaitForEvent(CLEventList list, int index, boolean blockingWait) {
 
         PointerBuffer ids = list.IDs.duplicate().position(index);
-        
         int ret = blockingWait ? cl.clWaitForEvents(1, ids)
                                : cl.clEnqueueWaitForEvents(ID, 1, ids);
         if(ret != CL_SUCCESS) {
@@ -1407,7 +1407,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
      */
     public CLCommandQueue putTask(CLKernel kernel, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1444,9 +1444,9 @@ public class CLCommandQueue extends CLObject implements CLResource {
      * Calls {@native clEnqueueNDRangeKernel}.
      */
     public CLCommandQueue put1DRangeKernel(CLKernel kernel, long globalWorkOffset, long globalWorkSize, long localWorkSize, CLEventList condition, CLEventList events) {
-        PointerBuffer globWO = null;
-        PointerBuffer globWS = null;
-        PointerBuffer locWS = null;
+        NativeSizeBuffer globWO = null;
+        NativeSizeBuffer globWS = null;
+        NativeSizeBuffer locWS = null;
 
         if(globalWorkOffset != 0) {
             globWO = copy2NIO(ibA, globalWorkOffset);
@@ -1495,17 +1495,17 @@ public class CLCommandQueue extends CLObject implements CLResource {
     public CLCommandQueue put2DRangeKernel(CLKernel kernel, long globalWorkOffsetX, long globalWorkOffsetY,
                                                             long globalWorkSizeX, long globalWorkSizeY,
                                                             long localWorkSizeX, long localWorkSizeY, CLEventList condition, CLEventList events) {
-        PointerBuffer globalWorkOffset = null;
-        PointerBuffer globalWorkSize = null;
-        PointerBuffer localWorkSize = null;
+        NativeSizeBuffer globalWorkOffset = null;
+        NativeSizeBuffer globalWorkSize = null;
+        NativeSizeBuffer localWorkSize = null;
 
-        if(globalWorkOffsetX != 0 && globalWorkOffsetY != 0) {
+        if(globalWorkOffsetX != 0 || globalWorkOffsetY != 0) {
             globalWorkOffset = copy2NIO(ibA, globalWorkOffsetX, globalWorkOffsetY);
         }
-        if(globalWorkSizeX != 0 && globalWorkSizeY != 0) {
+        if(globalWorkSizeX != 0 || globalWorkSizeY != 0) {
             globalWorkSize = copy2NIO(ibB, globalWorkSizeX, globalWorkSizeY);
         }
-        if(localWorkSizeX != 0 && localWorkSizeY !=0) {
+        if(localWorkSizeX != 0 || localWorkSizeY != 0) {
             localWorkSize = copy2NIO(ibC, localWorkSizeX, localWorkSizeY);
         }
         this.putNDRangeKernel(kernel, 2, globalWorkOffset, globalWorkSize, localWorkSize, condition, events);
@@ -1515,7 +1515,57 @@ public class CLCommandQueue extends CLObject implements CLResource {
     /**
      * Calls {@native clEnqueueNDRangeKernel}.
      */
-    public CLCommandQueue putNDRangeKernel(CLKernel kernel, int workDimension, PointerBuffer globalWorkOffset, PointerBuffer globalWorkSize, PointerBuffer localWorkSize) {
+    public CLCommandQueue put3DRangeKernel(CLKernel kernel, long globalWorkOffsetX, long globalWorkOffsetY, long globalWorkOffsetZ,
+                                                            long globalWorkSizeX, long globalWorkSizeY, long globalWorkSizeZ,
+                                                            long localWorkSizeX, long localWorkSizeY, long localWorkSizeZ) {
+        this.put3DRangeKernel(kernel,
+                globalWorkOffsetX, globalWorkOffsetY, globalWorkOffsetZ,
+                globalWorkSizeX, globalWorkSizeY, globalWorkSizeZ,
+                localWorkSizeX, localWorkSizeY, localWorkSizeZ, null, null);
+
+        return this;
+    }
+
+    /**
+     * Calls {@native clEnqueueNDRangeKernel}.
+     */
+    public CLCommandQueue put3DRangeKernel(CLKernel kernel, long globalWorkOffsetX, long globalWorkOffsetY, long globalWorkOffsetZ,
+                                                            long globalWorkSizeX, long globalWorkSizeY, long globalWorkSizeZ,
+                                                            long localWorkSizeX, long localWorkSizeY, long localWorkSizeZ, CLEventList events) {
+        this.put3DRangeKernel(kernel,
+                globalWorkOffsetX, globalWorkOffsetY, globalWorkOffsetZ,
+                globalWorkSizeX, globalWorkSizeY, globalWorkSizeZ,
+                localWorkSizeX, localWorkSizeY, localWorkSizeZ, null, events);
+        return this;
+    }
+
+    /**
+     * Calls {@native clEnqueueNDRangeKernel}.
+     */
+    public CLCommandQueue put3DRangeKernel(CLKernel kernel, long globalWorkOffsetX, long globalWorkOffsetY, long globalWorkOffsetZ,
+                                                            long globalWorkSizeX, long globalWorkSizeY, long globalWorkSizeZ,
+                                                            long localWorkSizeX, long localWorkSizeY, long localWorkSizeZ, CLEventList condition, CLEventList events) {
+        NativeSizeBuffer globalWorkOffset = null;
+        NativeSizeBuffer globalWorkSize = null;
+        NativeSizeBuffer localWorkSize = null;
+
+        if(globalWorkOffsetX != 0 || globalWorkOffsetY != 0 || globalWorkOffsetZ != 0) {
+            globalWorkOffset = copy2NIO(ibA, globalWorkOffsetX, globalWorkOffsetY, globalWorkOffsetZ);
+        }
+        if(globalWorkSizeX != 0 || globalWorkSizeY != 0 || globalWorkSizeZ != 0) {
+            globalWorkSize = copy2NIO(ibB, globalWorkSizeX, globalWorkSizeY, globalWorkSizeZ);
+        }
+        if(localWorkSizeX != 0 || localWorkSizeY != 0 || localWorkSizeZ != 0) {
+            localWorkSize = copy2NIO(ibC, localWorkSizeX, localWorkSizeY, localWorkSizeZ);
+        }
+        this.putNDRangeKernel(kernel, 3, globalWorkOffset, globalWorkSize, localWorkSize, condition, events);
+        return this;
+    }
+
+    /**
+     * Calls {@native clEnqueueNDRangeKernel}.
+     */
+    public CLCommandQueue putNDRangeKernel(CLKernel kernel, int workDimension, NativeSizeBuffer globalWorkOffset, NativeSizeBuffer globalWorkSize, NativeSizeBuffer localWorkSize) {
         this.putNDRangeKernel(kernel, workDimension, globalWorkOffset, globalWorkSize, localWorkSize, null, null);
         return this;
     }
@@ -1523,7 +1573,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
     /**
      * Calls {@native clEnqueueNDRangeKernel}.
      */
-    public CLCommandQueue putNDRangeKernel(CLKernel kernel, int workDimension, PointerBuffer globalWorkOffset, PointerBuffer globalWorkSize, PointerBuffer localWorkSize, CLEventList events) {
+    public CLCommandQueue putNDRangeKernel(CLKernel kernel, int workDimension, NativeSizeBuffer globalWorkOffset, NativeSizeBuffer globalWorkSize, NativeSizeBuffer localWorkSize, CLEventList events) {
         this.putNDRangeKernel(kernel, workDimension, globalWorkOffset, globalWorkSize, localWorkSize, null, events);
         return this;
     }
@@ -1531,10 +1581,10 @@ public class CLCommandQueue extends CLObject implements CLResource {
     /**
      * Calls {@native clEnqueueNDRangeKernel}.
      */
-    public CLCommandQueue putNDRangeKernel(CLKernel kernel, int workDimension, PointerBuffer globalWorkOffset,
-            PointerBuffer globalWorkSize, PointerBuffer localWorkSize, CLEventList condition, CLEventList events) {
+    public CLCommandQueue putNDRangeKernel(CLKernel kernel, int workDimension, NativeSizeBuffer globalWorkOffset,
+            NativeSizeBuffer globalWorkSize, NativeSizeBuffer localWorkSize, CLEventList condition, CLEventList events) {
 
-        PointerBuffer conditionIDs = null;
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1567,7 +1617,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
     /**
      * Calls {@native clEnqueueAcquireGLObjects}.
      */
-    public CLCommandQueue putAcquireGLObject(long glObject) {
+    public CLCommandQueue putAcquireGLObject(CLGLObject glObject) {
         this.putAcquireGLObject(glObject, null, null);
         return this;
     }
@@ -1575,7 +1625,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
     /**
      * Calls {@native clEnqueueAcquireGLObjects}.
      */
-    public CLCommandQueue putAcquireGLObject(long glObject, CLEventList events) {
+    public CLCommandQueue putAcquireGLObject(CLGLObject glObject, CLEventList events) {
         this.putAcquireGLObject(glObject, null, events);
         return this;
     }
@@ -1583,9 +1633,33 @@ public class CLCommandQueue extends CLObject implements CLResource {
     /**
      * Calls {@native clEnqueueAcquireGLObjects}.
      */
-    public CLCommandQueue putAcquireGLObject(long glObject, CLEventList condition, CLEventList events) {
+    public CLCommandQueue putAcquireGLObject(CLGLObject glObject, CLEventList condition, CLEventList events) {
+        this.putAcquireGLObjects(copy2NIO(ibA, glObject.getID()), condition, events);
+        return this;
+    }
+
+    /**
+     * Calls {@native clEnqueueAcquireGLObjects}.
+     */
+    public CLCommandQueue putAcquireGLObjects(CLGLObject glObject1, CLGLObject glObject2, CLEventList condition, CLEventList events) {
+        this.putAcquireGLObjects(copy2NIO(ibA, glObject1.getID(), glObject2.getID()), condition, events);
+        return this;
+    }
 
-        PointerBuffer conditionIDs = null;
+    /**
+     * Calls {@native clEnqueueAcquireGLObjects}.
+     */
+    public CLCommandQueue putAcquireGLObjects(CLGLObject glObject1, CLGLObject glObject2, CLGLObject glObject3, CLEventList condition, CLEventList events) {
+        this.putAcquireGLObjects(copy2NIO(ibA, glObject1.getID(), glObject2.getID(), glObject3.getID()), condition, events);
+        return this;
+    }
+
+    /**
+     * Calls {@native clEnqueueAcquireGLObjects}.
+     */
+    public CLCommandQueue putAcquireGLObjects(NativeSizeBuffer glObjectIDs, CLEventList condition, CLEventList events) {
+
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1594,14 +1668,12 @@ public class CLCommandQueue extends CLObject implements CLResource {
 
         CLGLI xl = (CLGLI) cl;
 
-        PointerBuffer glObj = copy2NIO(ibA, glObject);
-        
-        int ret = xl.clEnqueueAcquireGLObjects(ID, 1, glObj,
+        int ret = xl.clEnqueueAcquireGLObjects(ID, glObjectIDs.remaining(), glObjectIDs,
                     conditions, conditionIDs,
                     events==null ? null : events.IDs);
 
         if(ret != CL_SUCCESS) {
-            throw newException(ret, "can not aquire GLObject: " + glObject + " with " + toStr(condition, events));
+            throw newException(ret, "can not aquire " + glObjectIDs + " with " + toStr(condition, events));
         }
 
         if(events != null) {
@@ -1614,7 +1686,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
     /**
      * Calls {@native clEnqueueReleaseGLObjects}.
      */
-    public CLCommandQueue putReleaseGLObject(long glObject) {
+    public CLCommandQueue putReleaseGLObject(CLGLObject glObject) {
         this.putReleaseGLObject(glObject, null);
         return this;
     }
@@ -1622,7 +1694,7 @@ public class CLCommandQueue extends CLObject implements CLResource {
     /**
      * Calls {@native clEnqueueReleaseGLObjects}.
      */
-    public CLCommandQueue putReleaseGLObject(long glObject, CLEventList events) {
+    public CLCommandQueue putReleaseGLObject(CLGLObject glObject, CLEventList events) {
         this.putReleaseGLObject(glObject, null, events);
         return this;
     }
@@ -1630,9 +1702,33 @@ public class CLCommandQueue extends CLObject implements CLResource {
     /**
      * Calls {@native clEnqueueReleaseGLObjects}.
      */
-    public CLCommandQueue putReleaseGLObject(long glObject, CLEventList condition, CLEventList events) {
+    public CLCommandQueue putReleaseGLObject(CLGLObject glObject, CLEventList condition, CLEventList events) {
+        this.putReleaseGLObjects(copy2NIO(ibA, glObject.getID()), condition, events);
+        return this;
+    }
 
-        PointerBuffer conditionIDs = null;
+    /**
+     * Calls {@native clEnqueueAcquireGLObjects}.
+     */
+    public CLCommandQueue putReleaseGLObjects(CLGLObject glObject1, CLGLObject glObject2, CLEventList condition, CLEventList events) {
+        this.putReleaseGLObjects(copy2NIO(ibA, glObject1.getID(), glObject2.getID()), condition, events);
+        return this;
+    }
+
+    /**
+     * Calls {@native clEnqueueAcquireGLObjects}.
+     */
+    public CLCommandQueue putReleaseGLObjects(CLGLObject glObject1, CLGLObject glObject2, CLGLObject glObject3, CLEventList condition, CLEventList events) {
+        this.putReleaseGLObjects(copy2NIO(ibA, glObject1.getID(), glObject2.getID(), glObject3.getID()), condition, events);
+        return this;
+    }
+
+    /**
+     * Calls {@native clEnqueueReleaseGLObjects}.
+     */
+    public CLCommandQueue putReleaseGLObjects(NativeSizeBuffer glObjectIDs, CLEventList condition, CLEventList events) {
+
+        NativeSizeBuffer conditionIDs = null;
         int conditions = 0;
         if(condition != null) {
             conditionIDs = condition.IDsView;
@@ -1641,14 +1737,12 @@ public class CLCommandQueue extends CLObject implements CLResource {
 
         CLGLI xl = (CLGLI) cl;
 
-        PointerBuffer glObj = copy2NIO(ibA, glObject);
-
-        int ret = xl.clEnqueueReleaseGLObjects(ID, 1, glObj,
+        int ret = xl.clEnqueueReleaseGLObjects(ID, glObjectIDs.remaining(), glObjectIDs,
                 conditions, conditionIDs,
                 events==null ? null : events.IDs);
 
         if(ret != CL_SUCCESS) {
-            throw newException(ret, "can not release GLObject: " + glObject + "with " + toStr(condition, events));
+            throw newException(ret, "can not release " + glObjectIDs + "with " + toStr(condition, events));
         }
 
         if(events != null) {
@@ -1698,27 +1792,19 @@ public class CLCommandQueue extends CLObject implements CLResource {
         }
     }
 
-    private static PointerBuffer copy2NIO(PointerBuffer buffer, long a) {
+    private static NativeSizeBuffer copy2NIO(NativeSizeBuffer buffer, long a) {
         return buffer.put(2, a).position(2);
     }
 
-//    private static PointerBuffer copy2NIO(PointerBuffer buffer, long a, long b) {
-//        return buffer.position(1).put(a).put(b).position(1);
-//    }
-//
-//    private static PointerBuffer copy2NIO(PointerBuffer buffer, long a, long b, long c) {
-//        return buffer.rewind().put(a).put(b).put(c).rewind();
-//    }
-
-    private static PointerBuffer copy2NIO(PointerBuffer buffer, long a, long b) {
+    private static NativeSizeBuffer copy2NIO(NativeSizeBuffer buffer, long a, long b) {
         return buffer.position(1).put(a).put(b).position(1);
     }
 
-    private static PointerBuffer copy2NIO(PointerBuffer buffer, long a, long b, long c) {
+    private static NativeSizeBuffer copy2NIO(NativeSizeBuffer buffer, long a, long b, long c) {
         return buffer.rewind().put(a).put(b).put(c).rewind();
     }
 
-    private static String toStr(PointerBuffer buffer) {
+    private static String toStr(NativeSizeBuffer buffer) {
         if(buffer == null) {
             return null;
         }
diff --git a/src/com/jogamp/opencl/CLContext.java b/src/com/jogamp/opencl/CLContext.java
index c8a847a2..850f6c0e 100644
--- a/src/com/jogamp/opencl/CLContext.java
+++ b/src/com/jogamp/opencl/CLContext.java
@@ -32,7 +32,7 @@ import com.jogamp.common.nio.Buffers;
 import com.jogamp.opencl.CLDevice.Type;
 import com.jogamp.opencl.CLSampler.AddressingMode;
 import com.jogamp.opencl.CLSampler.FilteringMode;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import com.jogamp.opencl.impl.CLImageFormatImpl;
 import java.io.BufferedReader;
 import java.io.IOException;
@@ -119,7 +119,7 @@ public class CLContext extends CLObject implements CLResource {
         
         if (devices == null) {
 
-            PointerBuffer deviceCount = PointerBuffer.allocateDirect(1);
+            NativeSizeBuffer deviceCount = NativeSizeBuffer.allocateDirect(1);
 
             int ret = cl.clGetContextInfo(ID, CL.CL_CONTEXT_DEVICES, 0, null, deviceCount);
             checkForError(ret, "can not enumerate devices");
@@ -170,7 +170,7 @@ public class CLContext extends CLObject implements CLResource {
 
         long type = toDeviceBitmap(deviceTypes);
 
-        PointerBuffer properties = setupContextProperties(platform);
+        NativeSizeBuffer properties = setupContextProperties(platform);
         ErrorDispatcher dispatcher = new ErrorDispatcher();
         return new CLContext(platform, createContextFromType(dispatcher, properties, type), dispatcher);
     }
@@ -188,7 +188,7 @@ public class CLContext extends CLObject implements CLResource {
 
         CLPlatform platform = devices[0].getPlatform();
 
-        PointerBuffer properties = setupContextProperties(platform);
+        NativeSizeBuffer properties = setupContextProperties(platform);
         ErrorDispatcher dispatcher = new ErrorDispatcher();
         CLContext context = new CLContext(platform, createContext(dispatcher, properties, devices), dispatcher);
         if(devices != null) {
@@ -199,7 +199,7 @@ public class CLContext extends CLObject implements CLResource {
         return context;
     }
 
-    protected static long createContextFromType(CLErrorHandler handler, PointerBuffer properties, long deviceType) {
+    protected static long createContextFromType(CLErrorHandler handler, NativeSizeBuffer properties, long deviceType) {
 
         IntBuffer status = newDirectIntBuffer(1);
         long context = CLPlatform.getLowLevelCLInterface().clCreateContextFromType(properties, deviceType, handler, status);
@@ -209,12 +209,12 @@ public class CLContext extends CLObject implements CLResource {
         return context;
     }
 
-    protected static long createContext(CLErrorHandler handler, PointerBuffer properties, CLDevice... devices) {
+    protected static long createContext(CLErrorHandler handler, NativeSizeBuffer properties, CLDevice... devices) {
 
         IntBuffer status = newDirectIntBuffer(1);
-        PointerBuffer pb = null;
+        NativeSizeBuffer pb = null;
         if(devices != null && devices.length != 0) {
-            pb = PointerBuffer.allocateDirect(devices.length);
+            pb = NativeSizeBuffer.allocateDirect(devices.length);
             for (int i = 0; i < devices.length; i++) {
                 CLDevice device = devices[i];
                 if(device == null) {
@@ -230,13 +230,13 @@ public class CLContext extends CLObject implements CLResource {
         return context;
     }
 
-    private static PointerBuffer setupContextProperties(CLPlatform platform) {
+    private static NativeSizeBuffer setupContextProperties(CLPlatform platform) {
 
         if(platform == null) {
             throw new RuntimeException("no OpenCL installation found");
         }
 
-        return PointerBuffer.allocateDirect(3).put(CL.CL_CONTEXT_PLATFORM)
+        return NativeSizeBuffer.allocateDirect(3).put(CL.CL_CONTEXT_PLATFORM)
                                               .put(platform.ID).put(0) // 0 terminated array
                                               .rewind();
     }
diff --git a/src/com/jogamp/opencl/CLDevice.java b/src/com/jogamp/opencl/CLDevice.java
index 1e9a94dd..0381038e 100644
--- a/src/com/jogamp/opencl/CLDevice.java
+++ b/src/com/jogamp/opencl/CLDevice.java
@@ -29,10 +29,8 @@
 package com.jogamp.opencl;
 
 import com.jogamp.opencl.util.CLUtil;
-import com.jogamp.common.nio.PointerBuffer;
-import com.jogamp.common.os.Platform;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import java.nio.Buffer;
-import java.nio.ByteBuffer;
 import java.nio.ByteOrder;
 import java.util.ArrayList;
 import java.util.Collections;
@@ -63,13 +61,13 @@ public final class CLDevice extends CLObject {
     CLDevice(CL cl, CLPlatform platform, long id) {
         super(cl, id);
         this.platform = platform;
-        this.deviceInfo = new CLDeviceInfoAccessor();
+        this.deviceInfo = new CLDeviceInfoAccessor(cl, id);
     }
 
     CLDevice(CLContext context, long id) {
         super(context, id);
         this.platform = context.getPlatform();
-        this.deviceInfo = new CLDeviceInfoAccessor();
+        this.deviceInfo = new CLDeviceInfoAccessor(context.getCL(), id);
     }
 
     public CLCommandQueue createCommandQueue() {
@@ -340,8 +338,8 @@ public final class CLDevice extends CLObject {
      */
     @CLProperty("CL_DEVICE_MAX_WORK_ITEM_SIZES")
     public int[] getMaxWorkItemSizes() {
-        int n = (int) deviceInfo.getLong(CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS);
-        return deviceInfo.getInts(n, CL_DEVICE_MAX_WORK_ITEM_SIZES);
+        int n = getMaxWorkItemDimensions();
+        return deviceInfo.getInts(CL_DEVICE_MAX_WORK_ITEM_SIZES, n);
     }
 
     /**
@@ -693,35 +691,23 @@ public final class CLDevice extends CLObject {
         return CLUtil.obtainDeviceProperties(this);
     }
 
-    private final class CLDeviceInfoAccessor extends CLInfoAccessor {
+    private final static class CLDeviceInfoAccessor extends CLInfoAccessor {
 
-        @Override
-        protected int getInfo(int name, long valueSize, Buffer value, PointerBuffer valueSizeRet) {
-            return cl.clGetDeviceInfo(ID, name, valueSize, value, valueSizeRet);
-        }
+        private final CL cl;
+        private final long ID;
 
-        private int[] getInts(int n, int key) {
-
-            ByteBuffer buffer = localBB.get();
-            int ret = getInfo(key, buffer.capacity(), buffer, null);
-            CLException.checkForError(ret, "error while asking device for infos");
-
-            int[] array = new int[n];
-            for(int i = 0; i < array.length; i++) {
-                if(Platform.is32Bit()) {
-                    array[i] = buffer.getInt();
-                }else{
-                    array[i] = (int)buffer.getLong();
-                }
-            }
-            buffer.rewind();
+        private CLDeviceInfoAccessor(CL cl, long id) {
+            this.cl = cl;
+            this.ID = id;
+        }
 
-            return array;
+        @Override
+        protected int getInfo(int name, long valueSize, Buffer value, NativeSizeBuffer valueSizeRet) {
+            return cl.clGetDeviceInfo(ID, name, valueSize, value, valueSizeRet);
         }
 
     }
 
-
     @Override
     public String toString() {
         return "CLDevice [id: " + ID
diff --git a/src/com/jogamp/opencl/CLEvent.java b/src/com/jogamp/opencl/CLEvent.java
index 936ffe20..876be3da 100644
--- a/src/com/jogamp/opencl/CLEvent.java
+++ b/src/com/jogamp/opencl/CLEvent.java
@@ -29,10 +29,8 @@
 package com.jogamp.opencl;
 
 import com.jogamp.opencl.impl.CLEventCallback;
-import java.util.List;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import java.nio.Buffer;
-import java.util.ArrayList;
 
 import static com.jogamp.opencl.CL.*;
 import static com.jogamp.opencl.CLException.*;
@@ -143,7 +141,7 @@ public class CLEvent extends CLObject implements CLResource {
     private class CLEventInfoAccessor extends CLInfoAccessor {
 
         @Override
-        protected int getInfo(int name, long valueSize, Buffer value, PointerBuffer valueSizeRet) {
+        protected int getInfo(int name, long valueSize, Buffer value, NativeSizeBuffer valueSizeRet) {
             return cl.clGetEventInfo(ID, name, valueSize, value, valueSizeRet);
         }
 
@@ -152,7 +150,7 @@ public class CLEvent extends CLObject implements CLResource {
     private class CLEventProfilingInfoAccessor extends CLInfoAccessor {
 
         @Override
-        protected int getInfo(int name, long valueSize, Buffer value, PointerBuffer valueSizeRet) {
+        protected int getInfo(int name, long valueSize, Buffer value, NativeSizeBuffer valueSizeRet) {
             return cl.clGetEventProfilingInfo(ID, name, valueSize, value, valueSizeRet);
         }
 
diff --git a/src/com/jogamp/opencl/CLEventList.java b/src/com/jogamp/opencl/CLEventList.java
index 1e882221..fe6378b7 100644
--- a/src/com/jogamp/opencl/CLEventList.java
+++ b/src/com/jogamp/opencl/CLEventList.java
@@ -30,7 +30,7 @@ package com.jogamp.opencl;
 
 import com.jogamp.common.AutoCloseable;
 import com.jogamp.common.nio.CachedBufferFactory;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import java.util.Iterator;
 
 /**
@@ -44,12 +44,12 @@ public final class CLEventList implements CLResource, AutoCloseable, Iterable<CL
     /**
      * stores event ids for fast access.
      */
-    final PointerBuffer IDs;
+    final NativeSizeBuffer IDs;
     
     /**
      * Points always to the first element of the id buffer.
      */
-    final PointerBuffer IDsView;
+    final NativeSizeBuffer IDsView;
     
     int size;
     
@@ -82,9 +82,9 @@ public final class CLEventList implements CLResource, AutoCloseable, Iterable<CL
         size = events.length;
     }
     
-    private PointerBuffer initIDBuffer(CachedBufferFactory factory, int size) {
+    private NativeSizeBuffer initIDBuffer(CachedBufferFactory factory, int size) {
         if(factory == null) {
-            return PointerBuffer.allocateDirect(size);
+            return NativeSizeBuffer.allocateDirect(size);
         }else{
             return PointerBuffer.wrap(factory.newDirectByteBuffer(size*PointerBuffer.ELEMENT_SIZE));
         }
diff --git a/src/com/jogamp/opencl/CLImage.java b/src/com/jogamp/opencl/CLImage.java
index 335dc0a3..01dd6ea2 100644
--- a/src/com/jogamp/opencl/CLImage.java
+++ b/src/com/jogamp/opencl/CLImage.java
@@ -28,7 +28,7 @@
 
 package com.jogamp.opencl;
 
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import java.nio.Buffer;
 
 import static com.jogamp.opencl.CL.*;
@@ -73,6 +73,7 @@ public abstract class CLImage<B extends Buffer> extends CLMemory<B>  {
      * Returns the size of each element of the image memory object given by image.
      * An element is made up of n channels. The value of n is given in {@link CLImageFormat} descriptor.
      */
+    @Override
     public int getElementSize() {
         return (int)imageInfo.getLong(CL_IMAGE_ELEMENT_SIZE);
     }
@@ -109,7 +110,7 @@ public abstract class CLImage<B extends Buffer> extends CLMemory<B>  {
             this.id = id;
         }
         @Override
-        public int getInfo(int name, long valueSize, Buffer value, PointerBuffer valueSizeRet) {
+        public int getInfo(int name, long valueSize, Buffer value, NativeSizeBuffer valueSizeRet) {
             return cl.clGetImageInfo(id, name, valueSize, value, valueSizeRet);
         }
     }
diff --git a/src/com/jogamp/opencl/CLInfoAccessor.java b/src/com/jogamp/opencl/CLInfoAccessor.java
index 08f8d8a2..08d7305e 100644
--- a/src/com/jogamp/opencl/CLInfoAccessor.java
+++ b/src/com/jogamp/opencl/CLInfoAccessor.java
@@ -28,12 +28,13 @@
 
 package com.jogamp.opencl;
 
-import com.jogamp.common.nio.Buffers;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
+import com.jogamp.common.os.Platform;
 import com.jogamp.opencl.util.CLUtil;
 import java.nio.Buffer;
 import java.nio.ByteBuffer;
 
+import static com.jogamp.common.nio.Buffers.*;
 import static com.jogamp.opencl.CLException.*;
 
 /**
@@ -43,26 +44,28 @@ import static com.jogamp.opencl.CLException.*;
  */
 abstract class CLInfoAccessor {
 
+    private static final int BB_SIZE = 512;
+
     protected final static ThreadLocal<ByteBuffer> localBB = new ThreadLocal<ByteBuffer>() {
 
         @Override
         protected ByteBuffer initialValue() {
-            return Buffers.newDirectByteBuffer(512);
+            return newDirectByteBuffer(BB_SIZE);
         }
 
     };
-    protected final static ThreadLocal<PointerBuffer> localPB = new ThreadLocal<PointerBuffer>() {
+    protected final static ThreadLocal<NativeSizeBuffer> localNSB = new ThreadLocal<NativeSizeBuffer>() {
 
         @Override
-        protected PointerBuffer initialValue() {
-            return PointerBuffer.allocateDirect(1);
+        protected NativeSizeBuffer initialValue() {
+            return NativeSizeBuffer.allocateDirect(1);
         }
 
     };
 
     public final long getLong(int key) {
 
-        ByteBuffer buffer = localBB.get();
+        ByteBuffer buffer = getBB(8).putLong(0, 0);
         int ret = getInfo(key, 8, buffer, null);
         checkForError(ret, "error while asking for info value");
 
@@ -71,12 +74,16 @@ abstract class CLInfoAccessor {
 
     public final String getString(int key) {
         
-        ByteBuffer buffer = localBB.get();
-        PointerBuffer sizeBuffer = localPB.get();
-        int ret = getInfo(key, buffer.capacity(), buffer, sizeBuffer);
+        NativeSizeBuffer sizeBuffer = getNSB();
+        int ret = getInfo(key, 0, null, sizeBuffer);
         checkForError(ret, "error while asking for info string");
 
         int clSize = (int)sizeBuffer.get(0);
+        ByteBuffer buffer = getBB(clSize);
+
+        ret = getInfo(key, buffer.capacity(), buffer, null);
+        checkForError(ret, "error while asking for info string");
+
         byte[] array = new byte[clSize];
         buffer.get(array).rewind();
 
@@ -84,7 +91,38 @@ abstract class CLInfoAccessor {
 
     }
 
-    protected abstract int getInfo(int name, long valueSize, Buffer value, PointerBuffer valueSizeRet);
+    public final int[] getInts(int key, int n) {
+
+        ByteBuffer buffer = getBB(n * (Platform.is32Bit()?4:8));
+        int ret = getInfo(key, buffer.capacity(), buffer, null);
+        checkForError(ret, "error while asking for info value");
+
+        int[] array = new int[n];
+        for(int i = 0; i < array.length; i++) {
+            if(Platform.is32Bit()) {
+                array[i] = buffer.getInt();
+            }else{
+                array[i] = (int)buffer.getLong();
+            }
+        }
+        buffer.rewind();
+
+        return array;
+    }
+
+    protected ByteBuffer getBB(int minCapacity) {
+        if(minCapacity > BB_SIZE) {
+            return newDirectByteBuffer(minCapacity);
+        }else{
+            return localBB.get();
+        }
+    }
+
+    protected NativeSizeBuffer getNSB() {
+        return localNSB.get();
+    }
+
+    protected abstract int getInfo(int name, long valueSize, Buffer value, NativeSizeBuffer valueSizeRet);
 
 
 }
diff --git a/src/com/jogamp/opencl/CLKernel.java b/src/com/jogamp/opencl/CLKernel.java
index 8e246e5e..4c898b4b 100644
--- a/src/com/jogamp/opencl/CLKernel.java
+++ b/src/com/jogamp/opencl/CLKernel.java
@@ -30,7 +30,7 @@ package com.jogamp.opencl;
 
 import com.jogamp.opencl.util.CLUtil;
 import com.jogamp.common.nio.Buffers;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import java.nio.Buffer;
 import java.nio.ByteBuffer;
 
@@ -62,28 +62,36 @@ public class CLKernel extends CLObject implements CLResource, Cloneable {
     private boolean force32BitArgs;
 
     CLKernel(CLProgram program, long id) {
+        this(program, null, id);
+    }
+
+    CLKernel(CLProgram program, String name, long id) {
         super(program.getContext(), id);
+
         this.program = program;
         this.buffer = Buffers.newDirectByteBuffer((is32Bit()?4:8)*3);
 
-        PointerBuffer size = PointerBuffer.allocateDirect(1);
-
-        // get function name
-        int ret = cl.clGetKernelInfo(ID, CL_KERNEL_FUNCTION_NAME, 0, null, size);
-        checkForError(ret, "error while asking for kernel function name");
+        if(name == null) {
+            // get function name
+            NativeSizeBuffer size = NativeSizeBuffer.wrap(buffer);
+            int ret = cl.clGetKernelInfo(ID, CL_KERNEL_FUNCTION_NAME, 0, null, size);
+            checkForError(ret, "error while asking for kernel function name");
 
-        ByteBuffer bb = Buffers.newDirectByteBuffer((int)size.get(0));
+            ByteBuffer bb = Buffers.newDirectByteBuffer((int)size.get(0));
 
-        ret = cl.clGetKernelInfo(ID, CL_KERNEL_FUNCTION_NAME, bb.capacity(), bb, null);
-        checkForError(ret, "error while asking for kernel function name");
-
-        this.name = CLUtil.clString2JavaString(bb, bb.capacity());
+            ret = cl.clGetKernelInfo(ID, CL_KERNEL_FUNCTION_NAME, bb.capacity(), bb, null);
+            checkForError(ret, "error while asking for kernel function name");
+            
+            this.name = CLUtil.clString2JavaString(bb, bb.capacity());
+        }else{
+            this.name = name;
+        }
 
         // get number of arguments
-        ret = cl.clGetKernelInfo(ID, CL_KERNEL_NUM_ARGS, bb.capacity(), bb, null);
+        int ret = cl.clGetKernelInfo(ID, CL_KERNEL_NUM_ARGS, buffer.capacity(), buffer, null);
         checkForError(ret, "error while asking for number of function arguments.");
 
-        numArgs = bb.getInt(0);
+        numArgs = buffer.getInt(0);
 
     }
 
@@ -93,32 +101,38 @@ public class CLKernel extends CLObject implements CLResource, Cloneable {
 //    }
     
     public CLKernel putArg(CLMemory<?> value) {
-        setArg(argIndex++, value);
+        setArg(argIndex, value);
+        argIndex++;
         return this;
     }
 
     public CLKernel putArg(int value) {
-        setArg(argIndex++, value);
+        setArg(argIndex, value);
+        argIndex++;
         return this;
     }
 
     public CLKernel putArg(long value) {
-        setArg(argIndex++, value);
+        setArg(argIndex, value);
+        argIndex++;
         return this;
     }
 
     public CLKernel putArg(float value) {
-        setArg(argIndex++, value);
+        setArg(argIndex, value);
+        argIndex++;
         return this;
     }
 
     public CLKernel putArg(double value) {
-        setArg(argIndex++, value);
+        setArg(argIndex, value);
+        argIndex++;
         return this;
     }
 
     public CLKernel putNullArg(int size) {
-        setNullArg(argIndex++, size);
+        setNullArg(argIndex, size);
+        argIndex++;
         return this;
     }
 
@@ -128,11 +142,21 @@ public class CLKernel extends CLObject implements CLResource, Cloneable {
         return this;
     }
 
+    /**
+     * Resets the argument index to 0.
+     */
     public CLKernel rewind() {
         argIndex = 0;
         return this;
     }
 
+    /**
+     * Returns the argument index used in the relative putArt(...) methods.
+     */
+    public int position() {
+        return argIndex;
+    }
+
 //    public CLKernel setArg(int argumentIndex, Buffer value) {
 //        setArgument(argumentIndex, CLMemory.sizeOfBufferElem(value)*value.capacity(), value);
 //        return this;
@@ -224,19 +248,19 @@ public class CLKernel extends CLObject implements CLResource, Cloneable {
     }
 
     private Buffer wrap(float value) {
-        return buffer.putFloat(value).rewind();
+        return buffer.putFloat(0, value);
     }
 
     private Buffer wrap(double value) {
-        return buffer.putDouble(value).rewind();
+        return buffer.putDouble(0, value);
     }
 
     private Buffer wrap(int value) {
-        return buffer.putInt(value).rewind();
+        return buffer.putInt(0, value);
     }
 
     private Buffer wrap(long value) {
-        return buffer.putLong(value).rewind();
+        return buffer.putLong(0, value);
     }
 
     /**
diff --git a/src/com/jogamp/opencl/CLMemory.java b/src/com/jogamp/opencl/CLMemory.java
index b78e6024..db64f8a5 100644
--- a/src/com/jogamp/opencl/CLMemory.java
+++ b/src/com/jogamp/opencl/CLMemory.java
@@ -30,7 +30,7 @@ package com.jogamp.opencl;
 
 import com.jogamp.opencl.gl.CLGLI;
 import com.jogamp.common.nio.Buffers;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import com.jogamp.opencl.impl.CLMemObjectDestructorCallback;
 import java.nio.Buffer;
 import java.nio.IntBuffer;
@@ -50,7 +50,7 @@ public abstract class CLMemory <B extends Buffer> extends CLObject implements CL
     
     B buffer;
     protected final int FLAGS;
-    protected final long size;
+    protected long size;
     
     // depends on the nio buffer type
     protected int elementSize;
@@ -65,11 +65,15 @@ public abstract class CLMemory <B extends Buffer> extends CLObject implements CL
         this.buffer = directBuffer;
         this.FLAGS = flags;
         this.size = size;
-        initElementSizes();
+        initElementSize();
+        initCLCapacity();
     }
 
-    private void initElementSizes() {
+    private void initElementSize() {
         this.elementSize = (buffer==null) ? 1 : Buffers.sizeOfBufferElem(buffer);
+    }
+
+    protected final void initCLCapacity() {
         this.clCapacity  = (int) (size / elementSize);
     }
 
@@ -118,7 +122,8 @@ public abstract class CLMemory <B extends Buffer> extends CLObject implements CL
                     +" but got " + buffer.getClass());
         }
         this.buffer = buffer;
-        initElementSizes();
+        initElementSize();
+        initCLCapacity();
         return this;
     }
 
diff --git a/src/com/jogamp/opencl/CLPlatform.java b/src/com/jogamp/opencl/CLPlatform.java
index f5c94aed..ee6a6bdf 100644
--- a/src/com/jogamp/opencl/CLPlatform.java
+++ b/src/com/jogamp/opencl/CLPlatform.java
@@ -30,10 +30,11 @@ package com.jogamp.opencl;
 
 import com.jogamp.common.nio.Buffers;
 import com.jogamp.common.os.DynamicLookupHelper;
+import java.nio.Buffer;
 import java.security.PrivilegedAction;
 import com.jogamp.common.JogampRuntimeException;
 import com.jogamp.common.os.NativeLibrary;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import com.jogamp.gluegen.runtime.FunctionAddressResolver;
 import com.jogamp.opencl.util.CLUtil;
 import com.jogamp.opencl.impl.CLImpl;
@@ -41,7 +42,6 @@ import com.jogamp.opencl.impl.CLProcAddressTable;
 import com.jogamp.opencl.util.Filter;
 import com.jogamp.opencl.util.JOCLVersion;
 
-import java.nio.ByteBuffer;
 import java.nio.IntBuffer;
 import java.util.ArrayList;
 import java.util.Collections;
@@ -106,9 +106,12 @@ public final class CLPlatform {
 
     private Set<String> extensions;
 
+    private final CLPlatformInfoAccessor info;
 
     private CLPlatform(long id) {
+        initialize();
         this.ID = id;
+        this.info = new CLPlatformInfoAccessor(id, cl);
         this.version = new CLVersion(getInfoString(CL_PLATFORM_VERSION));
     }
 
@@ -226,7 +229,7 @@ public final class CLPlatform {
         checkForError(ret, "can not enumerate platforms");
 
         // receive platform ids
-        PointerBuffer platformId = PointerBuffer.allocateDirect(ib.get(0));
+        NativeSizeBuffer platformId = NativeSizeBuffer.allocateDirect(ib.get(0));
         ret = cl.clGetPlatformIDs(platformId.capacity(), platformId, null);
         checkForError(ret, "can not enumerate platforms");
 
@@ -273,14 +276,15 @@ public final class CLPlatform {
         initialize();
 
         List<CLDevice> list = new ArrayList<CLDevice>();
+        
         for(int t = 0; t < types.length; t++) {
             CLDevice.Type type = types[t];
 
-            PointerBuffer deviceIDs = getDeviceIDs(type.TYPE); 
+            long[] deviceIDs = info.getDeviceIDs(type.TYPE);
 
             //add device to list
-            for (int n = 0; n < deviceIDs.capacity(); n++) {
-                list.add(new CLDevice(cl, this, deviceIDs.get(n)));
+            for (int n = 0; n < deviceIDs.length; n++) {
+                list.add(new CLDevice(cl, this, deviceIDs[n]));
             }
         }
 
@@ -296,11 +300,11 @@ public final class CLPlatform {
 
         List<CLDevice> list = new ArrayList<CLDevice>();
         
-        PointerBuffer deviceIDs = getDeviceIDs(CL_DEVICE_TYPE_ALL);
+        long[] deviceIDs = info.getDeviceIDs(CL_DEVICE_TYPE_ALL);
 
         //add device to list
-        for (int n = 0; n < deviceIDs.capacity(); n++) {
-            CLDevice device = new CLDevice(cl, this, deviceIDs.get(n));
+        for (int n = 0; n < deviceIDs.length; n++) {
+            CLDevice device = new CLDevice(cl, this, deviceIDs[n]);
             addIfAccepted(device, list, filters);
         }
 
@@ -308,30 +312,6 @@ public final class CLPlatform {
 
     }
 
-    private PointerBuffer getDeviceIDs(long type) {
-        
-        IntBuffer ib = Buffers.newDirectIntBuffer(1);
-        
-        //find all devices
-        int ret = cl.clGetDeviceIDs(ID, type, 0, null, ib);
-        
-        PointerBuffer deviceIDs = null;
-        
-        // return null rather than throwing an exception
-        if(ret == CL.CL_DEVICE_NOT_FOUND || ib.get(0) == 0) {
-            deviceIDs = PointerBuffer.allocate(0);
-        }else{
-            deviceIDs = PointerBuffer.allocateDirect(ib.get(0));
-            
-            checkForError(ret, "error while enumerating devices");
-
-            ret = cl.clGetDeviceIDs(ID, type, deviceIDs.capacity(), deviceIDs, null);
-            checkForError(ret, "error while enumerating devices");
-        }
-        
-        return deviceIDs;
-    }
-    
     private static <I> void addIfAccepted(I item, List<I> list, Filter<I>[] filters) {
         if(filters == null) {
             list.add(item);
@@ -510,22 +490,57 @@ public final class CLPlatform {
      * Returns a info string in exchange for a key (CL_PLATFORM_*).
      */
     public String getInfoString(int key) {
-        PointerBuffer size = PointerBuffer.allocateDirect(1);
-        // TODO use cache/query size
-        ByteBuffer bb = ByteBuffer.allocateDirect(512);
+        return info.getString(key);
+    }
 
-        int ret = cl.clGetPlatformInfo(ID, key, bb.capacity(), bb, size);
-        checkForError(ret, "can not receive info string");
+    private final static class CLPlatformInfoAccessor extends CLInfoAccessor {
+
+        private final long ID;
+        private final CL cl;
+
+        private CLPlatformInfoAccessor(long id, CL cl) {
+            this.ID = id;
+            this.cl = cl;
+        }
+
+        @Override
+        protected int getInfo(int name, long valueSize, Buffer value, NativeSizeBuffer valueSizeRet) {
+            return cl.clGetPlatformInfo(ID, name, valueSize, value, valueSizeRet);
+        }
+
+        public long[] getDeviceIDs(long type) {
+
+            IntBuffer buffer = getBB(4).asIntBuffer();
+            int ret = cl.clGetDeviceIDs(ID, type, 0, null, buffer);
+            int count = buffer.get(0);
+
+            // return an empty buffer rather than throwing an exception
+            if(ret == CL.CL_DEVICE_NOT_FOUND || count == 0) {
+                return new long[0];
+            }else{
+                checkForError(ret, "error while enumerating devices");
+
+                NativeSizeBuffer deviceIDs = NativeSizeBuffer.wrap(getBB(count*NativeSizeBuffer.elementSize()));
+                ret = cl.clGetDeviceIDs(ID, type, count, deviceIDs, null);
+                checkForError(ret, "error while enumerating devices");
+
+                long[] ids = new long[count];
+                for (int i = 0; i < ids.length; i++) {
+                    ids[i] = deviceIDs.get(i);
+                }
+                return ids;
+            }
+
+        }
 
-        return CLUtil.clString2JavaString(bb, (int)size.get(0));
     }
 
     @Override
     public String toString() {
-        return "CLPlatform [name:" + getName()
-                         +" vendor:"+getVendor()
-                         +" profile:"+getProfile()
-                         +" version:"+getVersion()+"]";
+        return "CLPlatform [name: " + getName()
+                         +", vendor: "+getVendor()
+                         +", profile: "+getProfile()
+                         +", version: "+getVersion()+"]";
     }
 
     @Override
diff --git a/src/com/jogamp/opencl/CLProgram.java b/src/com/jogamp/opencl/CLProgram.java
index 56804533..54d8cc12 100644
--- a/src/com/jogamp/opencl/CLProgram.java
+++ b/src/com/jogamp/opencl/CLProgram.java
@@ -32,6 +32,7 @@ import com.jogamp.common.nio.CachedBufferFactory;
 import com.jogamp.opencl.util.CLProgramConfiguration;
 import com.jogamp.opencl.util.CLUtil;
 import com.jogamp.common.os.Platform;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import com.jogamp.common.nio.PointerBuffer;
 import com.jogamp.opencl.impl.BuildProgramCallback;
 import com.jogamp.opencl.util.CLBuildListener;
@@ -77,7 +78,7 @@ public class CLProgram extends CLObject implements CLResource {
 
         IntBuffer status = newDirectIntBuffer(1);
         
-        PointerBuffer length = PointerBuffer.allocateDirect(1).put(0, src.length());
+        NativeSizeBuffer length = NativeSizeBuffer.allocateDirect(1).put(0, src.length());
         String[] srcArray = new String[] {src};
         
         // Create the program
@@ -105,9 +106,9 @@ public class CLProgram extends CLObject implements CLResource {
         int deviceCount = binaries.size();
         
         CachedBufferFactory bf = CachedBufferFactory.create(binarySize + pbSize*deviceCount*3 + 4, true);
-        PointerBuffer devices     = PointerBuffer.wrap(bf.newDirectByteBuffer(deviceCount*pbSize));
+        NativeSizeBuffer devices  = NativeSizeBuffer.wrap(bf.newDirectByteBuffer(deviceCount*pbSize));
         PointerBuffer codeBuffers = PointerBuffer.wrap(bf.newDirectByteBuffer(deviceCount*pbSize));
-        PointerBuffer lengths     = PointerBuffer.wrap(bf.newDirectByteBuffer(deviceCount*pbSize));
+        NativeSizeBuffer lengths  = NativeSizeBuffer.wrap(bf.newDirectByteBuffer(deviceCount*pbSize));
         
         int i = 0;
         for (Map.Entry<CLDevice, byte[]> entry : entries) {
@@ -164,7 +165,7 @@ public class CLProgram extends CLObject implements CLResource {
             return "";
         }
 
-        PointerBuffer size = PointerBuffer.allocateDirect(1);
+        NativeSizeBuffer size = NativeSizeBuffer.allocateDirect(1);
 
         int ret = cl.clGetProgramBuildInfo(ID, device.ID, flag, 0, null, size);
         if(ret != CL_SUCCESS) {
@@ -187,7 +188,7 @@ public class CLProgram extends CLObject implements CLResource {
             return "";
         }
 
-        PointerBuffer size = PointerBuffer.allocateDirect(1);
+        NativeSizeBuffer size = NativeSizeBuffer.allocateDirect(1);
 
         int ret = cl.clGetProgramInfo(ID, flag, 0, null, size);
         checkForError(ret, "on clGetProgramInfo");
@@ -328,10 +329,10 @@ public class CLProgram extends CLObject implements CLResource {
             releaseKernels();
         }
 
-        PointerBuffer deviceIDs = null;
+        NativeSizeBuffer deviceIDs = null;
         int count = 0;
         if(devices != null && devices.length != 0) {
-            deviceIDs = PointerBuffer.allocateDirect(devices.length);
+            deviceIDs = NativeSizeBuffer.allocateDirect(devices.length);
             for (int i = 0; i < devices.length; i++) {
                 deviceIDs.put(i, devices[i].ID);
             }
@@ -406,7 +407,7 @@ public class CLProgram extends CLObject implements CLResource {
             throw newException(err[0], "unable to create Kernel with name: "+kernelName);
         }
 
-        CLKernel kernel = new CLKernel(this, id);
+        CLKernel kernel = new CLKernel(this, kernelName, id);
         kernels.add(kernel);
         return kernel;
     }
@@ -430,7 +431,7 @@ public class CLProgram extends CLObject implements CLResource {
 
         if(numKernels.get(0) > 0) {
 
-            PointerBuffer kernelIDs = PointerBuffer.allocateDirect(numKernels.get(0));
+            NativeSizeBuffer kernelIDs = NativeSizeBuffer.allocateDirect(numKernels.get(0));
             ret = cl.clCreateKernelsInProgram(ID, kernelIDs.capacity(), kernelIDs, null);
             if(ret != CL_SUCCESS) {
                 throw newException(ret, "can not create "+kernelIDs.capacity()+" kernels for "+this);
@@ -494,7 +495,7 @@ public class CLProgram extends CLObject implements CLResource {
         if(released) {
             return new CLDevice[0];
         }
-        PointerBuffer size = PointerBuffer.allocateDirect(1);
+        NativeSizeBuffer size = NativeSizeBuffer.allocateDirect(1);
         int ret = cl.clGetProgramInfo(ID, CL_PROGRAM_DEVICES, 0, null, size);
         if(ret != CL_SUCCESS) {
             throw newException(ret, "on clGetProgramInfo of "+this);
@@ -619,12 +620,13 @@ public class CLProgram extends CLObject implements CLResource {
 
         
         long address = InternalBufferUtil.getDirectBufferAddress(binaries);
-        PointerBuffer addresses = PointerBuffer.allocateDirect(sizes.capacity());
+        NativeSizeBuffer addresses = NativeSizeBuffer.allocateDirect(sizes.capacity());
         sizes.rewind();
         while(sizes.remaining() != 0) {
             addresses.put(address);
             address += sizes.get();
         }
+        addresses.rewind();
         
         ret = cl.clGetProgramInfo(ID, CL_PROGRAM_BINARIES, addresses.capacity()*addresses.elementSize(), addresses.getBuffer(), null);
         if(ret != CL_SUCCESS) {
diff --git a/src/com/jogamp/opencl/CLProgramBuilder.java b/src/com/jogamp/opencl/CLProgramBuilder.java
index 389adce8..e80b9923 100644
--- a/src/com/jogamp/opencl/CLProgramBuilder.java
+++ b/src/com/jogamp/opencl/CLProgramBuilder.java
@@ -37,6 +37,7 @@ import java.io.ObjectOutputStream;
 import java.io.Serializable;
 import java.util.ArrayList;
 import java.util.Arrays;
+import java.util.Collections;
 import java.util.Iterator;
 import java.util.LinkedHashMap;
 import java.util.LinkedHashSet;
@@ -167,9 +168,7 @@ public final class CLProgramBuilder implements CLProgramConfiguration, Serializa
 
     @Override
     public CLProgramBuilder withOptions(String... options) {
-        for (String option : options) {
-            optionSet.add(option);
-        }
+        optionSet.addAll(Arrays.asList(options));
         return this;
     }
 
@@ -272,27 +271,15 @@ public final class CLProgramBuilder implements CLProgramConfiguration, Serializa
         optionSet.clear();
         return this;
     }
-    
-    private int indexOf(CLDevice device, CLDevice[] devices) {
-        for (int i = 0; i < devices.length; i++) {
-            if(device.equals(devices[i])) {
-                return i;
-            }
-        }
-        return -1;
-    }
 
-    // format: { platform_suffix, num_binaries, (device.ID, length, binaries)+ }
+    // format: { platform_suffix, num_binaries, (device_name, length, binaries)+ }
     private void writeObject(ObjectOutputStream out) throws IOException {
         out.defaultWriteObject();
 
-        CLDevice[] deviceList = null;
         String suffix = null;
-        
+
         if(!binariesMap.isEmpty()) {
             CLPlatform platform = binariesMap.keySet().iterator().next().getPlatform();
-            deviceList = platform.listCLDevices();
-
             suffix = platform.getICDSuffix();
         }
         
@@ -303,12 +290,7 @@ public final class CLProgramBuilder implements CLProgramConfiguration, Serializa
             CLDevice device = entry.getKey();
             byte[] binaries = entry.getValue();
             
-            // we use the device index as identifier since there is currently no other way
-            // to distinguish identical devices via CL.
-            // it should be persistent between runs but may change on driver/hardware update. In this situations we would
-            // have to build from source anyway (build failures).
-            int index = indexOf(device, deviceList);
-            out.writeInt(index);
+            out.writeUTF(device.getName());
             out.writeInt(binaries.length);
             out.write(binaries);
         }
@@ -328,22 +310,28 @@ public final class CLProgramBuilder implements CLProgramConfiguration, Serializa
         
         this.binariesMap = new LinkedHashMap<CLDevice, byte[]>();
         
-        CLDevice[] devices = null;
+        List<CLDevice> devices;
         if(platform != null) {
-            devices = platform.listCLDevices();
+            devices = new ArrayList(Arrays.asList(platform.listCLDevices()));
+        }else{
+            devices = Collections.EMPTY_LIST;
         }
         
         int mapSize = in.readInt();
 
         for (int i = 0; i < mapSize; i++) {
-            int index = in.readInt();
+            String name = in.readUTF();
             int length = in.readInt();
             byte[] binaries = new byte[length];
             in.readFully(binaries);
             
-            // we ignore binaries we can't map to devices
-            if(devices != null && index >= 0 && index < devices.length) {
-                binariesMap.put(devices[index], binaries);
+            for (int d = 0; d < devices.size(); d++) {
+                CLDevice device = devices.get(d);
+                if(device.getName().equals(name)) {
+                    binariesMap.put(device, binaries);
+                    devices.remove(d);
+                    break;
+                }
             }
         }
     }
diff --git a/src/com/jogamp/opencl/CLSampler.java b/src/com/jogamp/opencl/CLSampler.java
index 73e153d9..f215cab1 100644
--- a/src/com/jogamp/opencl/CLSampler.java
+++ b/src/com/jogamp/opencl/CLSampler.java
@@ -28,7 +28,7 @@
 
 package com.jogamp.opencl;
 
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 
 import java.nio.Buffer;
 
@@ -84,7 +84,7 @@ public class CLSampler extends CLObject implements CLResource {
     private class CLSamplerInfoAccessor extends CLInfoAccessor {
 
         @Override
-        protected int getInfo(int name, long valueSize, Buffer value, PointerBuffer valueSizeRet) {
+        protected int getInfo(int name, long valueSize, Buffer value, NativeSizeBuffer valueSizeRet) {
             return cl.clGetSamplerInfo(ID, name, valueSize, value, valueSizeRet);
         }
 
diff --git a/src/com/jogamp/opencl/gl/CLGLBuffer.java b/src/com/jogamp/opencl/gl/CLGLBuffer.java
index 86fef680..148aa83d 100644
--- a/src/com/jogamp/opencl/gl/CLGLBuffer.java
+++ b/src/com/jogamp/opencl/gl/CLGLBuffer.java
@@ -28,9 +28,10 @@
 
 package com.jogamp.opencl.gl;
 
-import com.jogamp.opencl.CL;
 import com.jogamp.opencl.CLBuffer;
+import com.jogamp.opencl.CLCommandQueue;
 import com.jogamp.opencl.CLContext;
+import com.jogamp.opencl.CLException;
 
 import java.nio.Buffer;
 import javax.media.opengl.GLContext;
@@ -48,22 +49,22 @@ public final class CLGLBuffer<B extends Buffer> extends CLBuffer<B> implements C
      */
     public final int GLID;
 
-    private CLGLBuffer(CLContext context, B directBuffer, long id, int glObject, int flags) {
-        super(context, directBuffer, getSizeImpl(context.getCL(), id), id, flags);
+    private CLGLBuffer(CLContext context, B directBuffer, long id, int glObject, long size, int flags) {
+        super(context, directBuffer, size, id, flags);
         this.GLID = glObject;
     }
 
 
-    static <B extends Buffer> CLGLBuffer<B> create(CLContext context, B directBuffer, int flags, int glObject) {
+    static <B extends Buffer> CLGLBuffer<B> create(CLContext context, B directBuffer, long size, int flags, int glObject) {
         checkBuffer(directBuffer, flags);
         
-        CL cl = getCL(context);
-        int[] result = new int[1];
-        CLGLI clgli = (CLGLI)cl;
+        CLGLI clgli = (CLGLI)getCL(context);
         
+        int[] result = new int[1];
         long id = clgli.clCreateFromGLBuffer(context.ID, flags, glObject, result, 0);
+        CLException.checkForError(result[0], "can not create CLGLObject from #"+glObject);
 
-        return new CLGLBuffer<B>(context, directBuffer, id, glObject, flags);
+        return new CLGLBuffer<B>(context, directBuffer, id, glObject, size, flags);
     }
 
     static <B extends Buffer> void checkBuffer(B directBuffer, int flags) throws IllegalArgumentException {
@@ -75,6 +76,16 @@ public final class CLGLBuffer<B extends Buffer> extends CLBuffer<B> implements C
         }
     }
 
+    /**
+     * Updates the size of this CLGLBuffer by querying OpenGL.
+     * This method may only be called if this memory object has been acquired by calling
+     * {@link CLCommandQueue#putAcquireGLObject(com.jogamp.opencl.gl.CLGLObject)}.
+     */
+    public void updateSize() {
+        size = getSizeImpl(cl, ID);
+        initCLCapacity();
+    }
+
     public int getGLObjectID() {
         return GLID;
     }
@@ -94,7 +105,7 @@ public final class CLGLBuffer<B extends Buffer> extends CLBuffer<B> implements C
 
     @Override
     public <T extends Buffer> CLGLBuffer<T> cloneWith(T directBuffer) {
-        return new CLGLBuffer<T>(context, directBuffer, ID, GLID, FLAGS);
+        return new CLGLBuffer<T>(context, directBuffer, ID, GLID, size, FLAGS);
     }
 
     @Override
diff --git a/src/com/jogamp/opencl/gl/CLGLContext.java b/src/com/jogamp/opencl/gl/CLGLContext.java
index c220fcfc..ee6c2108 100644
--- a/src/com/jogamp/opencl/gl/CLGLContext.java
+++ b/src/com/jogamp/opencl/gl/CLGLContext.java
@@ -33,10 +33,11 @@ import com.jogamp.opencl.CLDevice;
 import java.nio.Buffer;
 import com.jogamp.opencl.CLMemory.Mem;
 import com.jogamp.opencl.CLPlatform;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import jogamp.opengl.GLContextImpl;
 import jogamp.opengl.egl.EGLContext;
 import jogamp.opengl.macosx.cgl.MacOSXCGLContext;
+import jogamp.opengl.macosx.cgl.CGL;
 import jogamp.opengl.windows.wgl.WindowsWGLContext;
 import jogamp.opengl.x11.glx.X11GLXContext;
 import javax.media.opengl.GLContext;
@@ -95,7 +96,7 @@ public final class CLGLContext extends CLContext {
         }
 
         long[] glID = new long[1];
-        PointerBuffer properties = setupContextProperties(platform, glContext, glID);
+        NativeSizeBuffer properties = setupContextProperties(platform, glContext, glID);
         ErrorDispatcher dispatcher = createErrorHandler();
         long clID = createContextFromType(dispatcher, properties, toDeviceBitmap(deviceTypes));
 
@@ -119,7 +120,7 @@ public final class CLGLContext extends CLContext {
         CLPlatform platform = devices[0].getPlatform();
 
         long[] glID = new long[1];
-        PointerBuffer properties = setupContextProperties(platform, glContext, glID);
+        NativeSizeBuffer properties = setupContextProperties(platform, glContext, glID);
         ErrorDispatcher dispatcher = createErrorHandler();
         long clID = createContext(dispatcher, properties, devices);
 
@@ -133,7 +134,7 @@ public final class CLGLContext extends CLContext {
     }
 
 
-    private static PointerBuffer setupContextProperties(CLPlatform platform, GLContext glContext, long[] glID) {
+    private static NativeSizeBuffer setupContextProperties(CLPlatform platform, GLContext glContext, long[] glID) {
 
         if(platform == null) {
             throw new RuntimeException("no OpenCL installation found");
@@ -151,14 +152,14 @@ public final class CLGLContext extends CLContext {
         GLContextImpl ctxImpl = (GLContextImpl)glContext;
         glID[0] = glContext.getHandle();
 
-        PointerBuffer properties;
+        NativeSizeBuffer properties;
         if(glContext instanceof X11GLXContext) {
 //          spec: "When the GLX binding API is supported, the attribute
 //          CL_GL_CONTEXT_KHR should be set to a GLXContext handle to an
 //          OpenGL context, and the attribute CL_GLX_DISPLAY_KHR should be
 //          set to the Display handle of the X Window System display used to
 //          create the OpenGL context."
-            properties = PointerBuffer.allocateDirect(7);
+            properties = NativeSizeBuffer.allocateDirect(7);
             long displayHandle = ctxImpl.getDrawableImpl().getNativeSurface().getDisplayHandle();
             properties.put(CL_GL_CONTEXT_KHR).put(glID[0])
                       .put(CL_GLX_DISPLAY_KHR).put(displayHandle)
@@ -168,18 +169,19 @@ public final class CLGLContext extends CLContext {
 //          CL_GL_CONTEXT_KHR should be set to an HGLRC handle to an OpenGL
 //          context, and the attribute CL_WGL_HDC_KHR should be set to the
 //          HDC handle of the display used to create the OpenGL context."
-            properties = PointerBuffer.allocateDirect(7);
+            properties = NativeSizeBuffer.allocateDirect(7);
             long surfaceHandle = ctxImpl.getDrawableImpl().getNativeSurface().getSurfaceHandle();
             properties.put(CL_GL_CONTEXT_KHR).put(glID[0])
                       .put(CL_WGL_HDC_KHR).put(surfaceHandle)
                       .put(CL_CONTEXT_PLATFORM).put(platform.ID);
         }else if(glContext instanceof MacOSXCGLContext) {
-//            TODO test on mac
 //          spec: "When the CGL binding API is supported, the attribute
 //          CL_CGL_SHAREGROUP_KHR should be set to a CGLShareGroup handle to
 //          a CGL share group object."
-            properties = PointerBuffer.allocateDirect(5);
-            properties.put(CL_CGL_SHAREGROUP_KHR).put(glID[0])
+            long cgl = CGL.getCGLContext(glID[0]);
+            long group = CGL.CGLGetShareGroup(cgl);
+            properties = NativeSizeBuffer.allocateDirect(5);
+            properties.put(CL_CGL_SHAREGROUP_KHR).put(group)
                       .put(CL_CONTEXT_PLATFORM).put(platform.ID);
         }else if(glContext instanceof EGLContext) {
 //            TODO test EGL
@@ -188,7 +190,7 @@ public final class CLGLContext extends CLContext {
 //          OpenGL ES or OpenGL context, and the attribute
 //          CL_EGL_DISPLAY_KHR should be set to the EGLDisplay handle of the
 //          display used to create the OpenGL ES or OpenGL context."
-            properties = PointerBuffer.allocateDirect(7);
+            properties = NativeSizeBuffer.allocateDirect(7);
             long displayHandle = ctxImpl.getDrawableImpl().getNativeSurface().getDisplayHandle();
             properties.put(CL_GL_CONTEXT_KHR).put(glID[0])
                       .put(CL_EGL_DISPLAY_KHR).put(displayHandle)
@@ -201,20 +203,46 @@ public final class CLGLContext extends CLContext {
     }
 
     // Buffers
-    public final CLGLBuffer<?> createFromGLBuffer(int glBuffer, Mem... flags) {
-        return createFromGLBuffer(null, glBuffer, Mem.flagsToInt(flags));
+    /**
+     * Creates a CLGLBuffer for memory sharing with the specified OpenGL buffer.
+     * @param glBuffer The OpenGL buffer handle like a vertex buffer or pixel buffer object.
+     * @param glBufferSize The size of the OpenGL buffer in bytes
+     * @param flags optional flags.
+     */
+    public final CLGLBuffer<?> createFromGLBuffer(int glBuffer, long glBufferSize, Mem... flags) {
+        return createFromGLBuffer(null, glBuffer, glBufferSize, Mem.flagsToInt(flags));
     }
 
-    public final CLGLBuffer<?> createFromGLBuffer(int glBuffer, int flags) {
-        return createFromGLBuffer(null, glBuffer, flags);
+    /**
+     * Creates a CLGLBuffer for memory sharing with the specified OpenGL buffer.
+     * @param glBuffer The OpenGL buffer handle like a vertex buffer or pixel buffer object.
+     * @param glBufferSize The size of the OpenGL buffer in bytes
+     * @param flags optional flags.
+     */
+    public final CLGLBuffer<?> createFromGLBuffer(int glBuffer, long glBufferSize, int flags) {
+        return createFromGLBuffer(null, glBuffer, glBufferSize, flags);
     }
 
-    public final <B extends Buffer> CLGLBuffer<B> createFromGLBuffer(B directBuffer, int glBuffer, Mem... flags) {
-        return createFromGLBuffer(directBuffer, glBuffer, Mem.flagsToInt(flags));
+    /**
+     * Creates a CLGLBuffer for memory sharing with the specified OpenGL buffer.
+     * @param directBuffer A direct allocated NIO buffer for data transfers between java and OpenCL.
+     * @param glBuffer The OpenGL buffer handle like a vertex buffer or pixel buffer object.
+     * @param glBufferSize The size of the OpenGL buffer in bytes
+     * @param flags optional flags.
+     */
+    public final <B extends Buffer> CLGLBuffer<B> createFromGLBuffer(B directBuffer, int glBuffer, long glBufferSize, Mem... flags) {
+        return createFromGLBuffer(directBuffer, glBuffer, glBufferSize, Mem.flagsToInt(flags));
     }
 
-    public final <B extends Buffer> CLGLBuffer<B> createFromGLBuffer(B directBuffer, int glBuffer, int flags) {
-        CLGLBuffer<B> buffer = CLGLBuffer.create(this, directBuffer, flags, glBuffer);
+    /**
+     * Creates a CLGLBuffer for memory sharing with the specified OpenGL buffer.
+     * @param directBuffer A direct allocated NIO buffer for data transfers between java and OpenCL.
+     * @param glBuffer The OpenGL buffer handle like a vertex buffer or pixel buffer object.
+     * @param glBufferSize The size of the OpenGL buffer in bytes
+     * @param flags optional flags.
+     */
+    public final <B extends Buffer> CLGLBuffer<B> createFromGLBuffer(B directBuffer, int glBuffer, long glBufferSize, int flags) {
+        CLGLBuffer<B> buffer = CLGLBuffer.create(this, directBuffer, glBufferSize, flags, glBuffer);
         memoryObjects.add(buffer);
         return buffer;
     }
diff --git a/src/com/jogamp/opencl/gl/CLGLObject.java b/src/com/jogamp/opencl/gl/CLGLObject.java
index f0081ce5..a9bad53d 100644
--- a/src/com/jogamp/opencl/gl/CLGLObject.java
+++ b/src/com/jogamp/opencl/gl/CLGLObject.java
@@ -38,7 +38,7 @@ import javax.media.opengl.GLContext;
  *
  * @author Michael Bien
  */
-interface CLGLObject {
+public interface CLGLObject {
 
     /**
      * Returns the OpenGL object id of this shared object.
@@ -46,6 +46,11 @@ interface CLGLObject {
     public int getGLObjectID();
 
     /**
+     * Returns the OpenCL object handle
+     */
+    public long getID();
+
+    /**
      * Returns the OpenGL buffer type of this shared object.
      */
     public GLObjectType getGLObjectType();
diff --git a/src/com/jogamp/opencl/impl/CLImpl.java b/src/com/jogamp/opencl/impl/CLImpl.java
index 8945d32c..8ea44201 100644
--- a/src/com/jogamp/opencl/impl/CLImpl.java
+++ b/src/com/jogamp/opencl/impl/CLImpl.java
@@ -31,7 +31,7 @@
  */
 package com.jogamp.opencl.impl;
 
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 import com.jogamp.common.os.Platform;
 import com.jogamp.common.util.LongLongHashMap;
 import com.jogamp.opencl.CLErrorHandler;
@@ -56,7 +56,7 @@ public class CLImpl extends CLAbstractImpl {
         this.contextCallbackMap.setKeyNotFoundValue(0);
     }
 
-    public long clCreateContext(PointerBuffer properties, PointerBuffer devices, CLErrorHandler pfn_notify, IntBuffer errcode_ret) {
+    public long clCreateContext(NativeSizeBuffer properties, NativeSizeBuffer devices, CLErrorHandler pfn_notify, IntBuffer errcode_ret) {
 
         if (properties != null && !properties.isDirect()) {
             throw new RuntimeException("Argument \"properties\" was not a direct buffer");
@@ -87,7 +87,7 @@ public class CLImpl extends CLAbstractImpl {
 
     private native long clCreateContext0(Object cl_context_properties, int props_offset, int numDevices, Object devices, int devices_offset, Object pfn_notify, long[] global, Object errcode_ret, int err_offset, long address);
 
-    public long clCreateContextFromType(PointerBuffer properties, long device_type, CLErrorHandler pfn_notify, IntBuffer errcode_ret) {
+    public long clCreateContextFromType(NativeSizeBuffer properties, long device_type, CLErrorHandler pfn_notify, IntBuffer errcode_ret) {
 
         if (properties != null && !properties.isDirect()) {
             throw new RuntimeException("Argument \"properties\" was not a direct buffer");
@@ -134,7 +134,7 @@ public class CLImpl extends CLAbstractImpl {
     public native int clReleaseContextImpl(long context, long global, long address);
 
     /** Interface to C language function: <br> <code> int32_t clBuildProgram(cl_program, uint32_t, cl_device_id * , const char * , void * ); </code>    */
-    public int clBuildProgram(long program, int deviceCount, PointerBuffer deviceList, String options, BuildProgramCallback cb) {
+    public int clBuildProgram(long program, int deviceCount, NativeSizeBuffer deviceList, String options, BuildProgramCallback cb) {
 
         if (deviceList != null && !deviceList.isDirect()) {
             throw new RuntimeException("Argument \"properties\" was not a direct buffer");
@@ -175,38 +175,38 @@ public class CLImpl extends CLAbstractImpl {
 
 
     /** Interface to C language function: <br> <code> void *  {@native clEnqueueMapImage}(cl_command_queue command_queue, cl_mem image, uint32_t blocking_map, uint64_t map_flags, const size_t * , const size_t * , size_t *  image_row_pitch, size_t *  image_slice_pitch, uint32_t num_events_in_wait_list, cl_event *  event_wait_list, cl_event *  event, int32_t *  errcode_ret); </code>
-    @param origin a direct {@link com.jogamp.common.nio.PointerBuffer}
-    @param range a direct {@link com.jogamp.common.nio.PointerBuffer}
-    @param image_row_pitch a direct {@link com.jogamp.common.nio.PointerBuffer}
-    @param image_slice_pitch a direct {@link com.jogamp.common.nio.PointerBuffer}
-    @param event_wait_list a direct {@link com.jogamp.common.nio.PointerBuffer}
-    @param event a direct {@link com.jogamp.common.nio.PointerBuffer}
+    @param origin a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param range a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param image_row_pitch a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param image_slice_pitch a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param event_wait_list a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param event a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
     @param errcode_ret a direct {@link java.nio.IntBuffer}   */
-    public java.nio.ByteBuffer clEnqueueMapImage(long command_queue, long image, int blocking_map, long map_flags,
-            PointerBuffer origin, PointerBuffer range,
-            PointerBuffer image_row_pitch, PointerBuffer image_slice_pitch,
+    public ByteBuffer clEnqueueMapImage(long command_queue, long image, int blocking_map, long map_flags,
+            NativeSizeBuffer origin, NativeSizeBuffer range,
+            NativeSizeBuffer image_row_pitch, NativeSizeBuffer image_slice_pitch,
             int num_events_in_wait_list,
-            PointerBuffer event_wait_list, PointerBuffer event, java.nio.IntBuffer errcode_ret) {
+            NativeSizeBuffer event_wait_list, NativeSizeBuffer event, IntBuffer errcode_ret) {
 
-        if (!isDirect(origin)) {
+        if (origin != null && !origin.isDirect()) {
             throw new CLException("Argument \"origin\" was not a direct buffer");
         }
-        if (!isDirect(range)) {
+        if (range != null && !range.isDirect()) {
             throw new CLException("Argument \"range\" was not a direct buffer");
         }
-        if (!isDirect(image_row_pitch)) {
+        if (image_row_pitch != null && !image_row_pitch.isDirect()) {
             throw new CLException("Argument \"image_row_pitch\" was not a direct buffer");
         }
-        if (!isDirect(image_slice_pitch)) {
+        if (image_slice_pitch != null && !image_slice_pitch.isDirect()) {
             throw new CLException("Argument \"image_slice_pitch\" was not a direct buffer");
         }
-        if (!isDirect(event_wait_list)) {
+        if (event_wait_list != null && !event_wait_list.isDirect()) {
             throw new CLException("Argument \"event_wait_list\" was not a direct buffer");
         }
-        if (!isDirect(event)) {
+        if (event != null && !event.isDirect()) {
             throw new CLException("Argument \"event\" was not a direct buffer");
         }
-        if (!isDirect(errcode_ret)) {
+        if (errcode_ret != null && !errcode_ret.isDirect()) {
             throw new CLException("Argument \"errcode_ret\" was not a direct buffer");
         }
 
@@ -218,7 +218,7 @@ public class CLImpl extends CLAbstractImpl {
         if (getImageInfoAddress == 0) {
             throw new UnsupportedOperationException("Method not available");
         }
-        java.nio.ByteBuffer _res;
+        ByteBuffer _res;
         _res = clEnqueueMapImage0(command_queue, image, blocking_map, map_flags, origin != null ? origin.getBuffer() : null,
                 getDirectBufferByteOffset(origin), range != null ? range.getBuffer() : null,
                 getDirectBufferByteOffset(range), image_row_pitch != null ? image_row_pitch.getBuffer() : null,
@@ -235,12 +235,12 @@ public class CLImpl extends CLAbstractImpl {
     }
 
     /** Entry point to C language function: <code> void *  {@native clEnqueueMapImage}(cl_command_queue command_queue, cl_mem image, uint32_t blocking_map, uint64_t map_flags, const size_t * , const size_t * , size_t *  image_row_pitch, size_t *  image_slice_pitch, uint32_t num_events_in_wait_list, cl_event *  event_wait_list, cl_event *  event, int32_t *  errcode_ret); </code>
-    @param origin a direct {@link com.jogamp.gluegen.runtime.PointerBuffer}
-    @param range a direct {@link com.jogamp.gluegen.runtime.PointerBuffer}
-    @param image_row_pitch a direct {@link com.jogamp.gluegen.runtime.PointerBuffer}
-    @param image_slice_pitch a direct {@link com.jogamp.gluegen.runtime.PointerBuffer}
-    @param event_wait_list a direct {@link com.jogamp.gluegen.runtime.PointerBuffer}
-    @param event a direct {@link com.jogamp.gluegen.runtime.PointerBuffer}
+    @param origin a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param range a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param image_row_pitch a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param image_slice_pitch a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param event_wait_list a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
+    @param event a direct {@link com.jogamp.gluegen.common.nio.NativeSizeBuffer}
     @param errcode_ret a direct {@link java.nio.IntBuffer}   */
     private native ByteBuffer clEnqueueMapImage0(long command_queue, long image, int blocking_map, long map_flags,
             Object origin, int origin_byte_offset, Object range, int range_byte_offset, Object image_row_pitch,
diff --git a/src/com/jogamp/opencl/util/CLBuildListener.java b/src/com/jogamp/opencl/util/CLBuildListener.java
index 7e6dfc0b..9a520582 100644
--- a/src/com/jogamp/opencl/util/CLBuildListener.java
+++ b/src/com/jogamp/opencl/util/CLBuildListener.java
@@ -40,7 +40,7 @@ import com.jogamp.opencl.CLProgram;
  * Note1: registering a build callback can make {@link com.jogamp.opencl.CL#clBuildProgram} non blocking (OpenCL implementation dependent).<br/>
  * Note2: the thread which calls this method is unspecified. The Application should ensure propper synchronization.
  * @author Michael Bien
- * @see com.jogamp.opencl.CL#clBuildProgram(long, int, com.jogamp.common.nio.PointerBuffer, java.lang.String, com.jogamp.opencl.impl.BuildProgramCallback) 
+ * @see com.jogamp.opencl.CL#clBuildProgram(long, int, com.jogamp.common.nio.NativeSizeBuffer, java.lang.String, com.jogamp.opencl.impl.BuildProgramCallback)
  */
 public interface CLBuildListener {
 
diff --git a/src/com/jogamp/opencl/util/CLDeviceFilters.java b/src/com/jogamp/opencl/util/CLDeviceFilters.java
index a2ba0475..045d4c7f 100644
--- a/src/com/jogamp/opencl/util/CLDeviceFilters.java
+++ b/src/com/jogamp/opencl/util/CLDeviceFilters.java
@@ -32,6 +32,7 @@ import com.jogamp.opencl.CLCommandQueue.Mode;
 import com.jogamp.opencl.CLDevice;
 import java.nio.ByteOrder;
 import java.util.Arrays;
+import java.util.List;
 
 /**
  * Pre-defined filters.
@@ -82,8 +83,9 @@ public class CLDeviceFilters {
      */
     public static Filter<CLDevice> extension(final String... extensions) {
         return new Filter<CLDevice>() {
+            private final List<String> extensionList = Arrays.asList(extensions);
             public boolean accept(CLDevice item) {
-                return item.getExtensions().containsAll(Arrays.asList(extensions));
+                return item.getExtensions().containsAll(extensionList);
             }
         };
     }
@@ -93,8 +95,9 @@ public class CLDeviceFilters {
      */
     public static Filter<CLDevice> queueMode(final Mode... modes) {
         return new Filter<CLDevice>() {
+            private final List<Mode> modeList = Arrays.asList(modes);
             public boolean accept(CLDevice item) {
-                return item.getQueueProperties().containsAll(Arrays.asList(modes));
+                return item.getQueueProperties().containsAll(modeList);
             }
         };
     }
diff --git a/src/com/jogamp/opencl/util/CLMultiContext.java b/src/com/jogamp/opencl/util/CLMultiContext.java
new file mode 100644
index 00000000..f74c0a35
--- /dev/null
+++ b/src/com/jogamp/opencl/util/CLMultiContext.java
@@ -0,0 +1,158 @@
+/*
+ * Created on Thursday, April 28 2011 22:10
+ */
+package com.jogamp.opencl.util;
+
+import com.jogamp.opencl.CLContext;
+import com.jogamp.opencl.CLDevice;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.CLResource;
+import java.util.ArrayList;
+import java.util.Collection;
+import java.util.Collections;
+import java.util.HashMap;
+import java.util.List;
+import java.util.Map;
+
+import static java.util.Arrays.*;
+import static com.jogamp.opencl.CLDevice.Type.*;
+
+/**
+ * Utility for organizing multiple {@link CLContext}s.
+ *
+ * @author Michael Bien
+ */
+public class CLMultiContext implements CLResource {
+
+    private final List<CLContext> contexts;
+
+    private CLMultiContext() {
+        contexts = new ArrayList<CLContext>();
+    }
+
+    /**
+     * Creates a multi context with all devices of the specified platforms.
+     */
+    public static CLMultiContext create(CLPlatform... platforms) {
+        return create(platforms, ALL);
+    }
+
+    /**
+     * Creates a multi context with all devices of the specified platforms and types.
+     */
+    public static CLMultiContext create(CLPlatform[] platforms, CLDevice.Type... types) {
+        
+        if(platforms == null) {
+            throw new NullPointerException("platform list was null");
+        }else if(platforms.length == 0) {
+            throw new IllegalArgumentException("platform list was empty");
+        }
+
+        List<CLDevice> devices = new ArrayList<CLDevice>();
+        for (CLPlatform platform : platforms) {
+            devices.addAll(asList(platform.listCLDevices(types)));
+        }
+        return create(devices);
+    }
+
+    /**
+     * Creates a multi context with the specified devices.
+     * The devices don't have to be from the same platform.
+     */
+    public static CLMultiContext create(Collection<CLDevice> devices) {
+
+        if(devices.isEmpty()) {
+            throw new IllegalArgumentException("device list was empty");
+        }
+
+        Map<CLPlatform, List<CLDevice>> platformDevicesMap = filterPlatformConflicts(devices);
+
+        // create contexts
+        CLMultiContext mc = new CLMultiContext();
+        for (Map.Entry<CLPlatform, List<CLDevice>> entry : platformDevicesMap.entrySet()) {
+            List<CLDevice> list = entry.getValue();
+            // one context per device to workaround driver bugs
+            for (CLDevice device : list) {
+                CLContext context = CLContext.create(device);
+                mc.contexts.add(context);
+            }
+        }
+
+        return mc;
+    }
+
+    /**
+     * Creates a multi context with specified contexts.
+     */
+    public static CLMultiContext wrap(CLContext... contexts) {
+        CLMultiContext mc = new CLMultiContext();
+        mc.contexts.addAll(asList(contexts));
+        return mc;
+    }
+
+    /**
+     * filter devices; don't allow the same device to be used in more than one platform.
+     * example: a CPU available via the AMD and Intel SDKs shouldn't end up in two contexts
+     */
+    private static Map<CLPlatform, List<CLDevice>> filterPlatformConflicts(Collection<CLDevice> devices) {
+
+        // FIXME: devicename-platform is used as unique device identifier - replace if we have something better
+        
+        Map<CLPlatform, List<CLDevice>> filtered = new HashMap<CLPlatform, List<CLDevice>>();
+        Map<String, CLPlatform> used = new HashMap<String, CLPlatform>();
+
+        for (CLDevice device : devices) {
+
+            String name = device.getName(); 
+
+            CLPlatform platform = device.getPlatform();
+            CLPlatform usedPlatform = used.get(name);
+
+            if(usedPlatform == null || platform.equals(usedPlatform)) {
+                if(!filtered.containsKey(platform)) {
+                    filtered.put(platform, new ArrayList<CLDevice>());
+                }
+                filtered.get(platform).add(device);
+                used.put(name, platform);
+            }
+            
+        }
+        return filtered;
+    }
+
+
+    /**
+     * Releases all contexts.
+     * @see CLContext#release()
+     */
+    public void release() {
+        for (CLContext context : contexts) {
+            context.release();
+        }
+        contexts.clear();
+    }
+
+    public List<CLContext> getContexts() {
+        return Collections.unmodifiableList(contexts);
+    }
+
+    /**
+     * Returns a list containing all devices used in this multi context.
+     */
+    public List<CLDevice> getDevices() {
+        List<CLDevice> devices = new ArrayList<CLDevice>();
+        for (CLContext context : contexts) {
+            devices.addAll(asList(context.getDevices()));
+        }
+        return devices;
+    }
+
+    @Override
+    public String toString() {
+        return getClass().getSimpleName()+" [" + contexts.size()+" contexts, "
+                                               + getDevices().size()+ " devices]";
+    }
+
+
+
+}
diff --git a/src/com/jogamp/opencl/util/CLPlatformFilters.java b/src/com/jogamp/opencl/util/CLPlatformFilters.java
index dab7448f..48d20916 100644
--- a/src/com/jogamp/opencl/util/CLPlatformFilters.java
+++ b/src/com/jogamp/opencl/util/CLPlatformFilters.java
@@ -33,7 +33,8 @@ import com.jogamp.opencl.CLDevice;
 import com.jogamp.opencl.CLPlatform;
 import com.jogamp.opencl.CLVersion;
 import java.util.Arrays;
-import java.util.List;
+import javax.media.opengl.GL;
+import javax.media.opengl.GLContext;
 
 /**
  * Pre-defined filters.
@@ -70,10 +71,11 @@ public class CLPlatformFilters {
      */
     public static Filter<CLPlatform> glSharing() {
         return new Filter<CLPlatform>() {
+            private final Filter<CLDevice> glFilter = CLDeviceFilters.glSharing();
             public boolean accept(CLPlatform item) {
                 CLDevice[] devices = item.listCLDevices();
                 for (CLDevice device : devices) {
-                    if(device.isGLMemorySharingSupported()) {
+                    if(glFilter.accept(device)) {
                         return true;
                     }
                 }
@@ -83,6 +85,21 @@ public class CLPlatformFilters {
     }
 
     /**
+     * Accepts all with the given OpenGL context compatible platforms containing at least one
+     * devices of which supports OpenGL-OpenCL interoparability.
+     */
+    public static Filter<CLPlatform> glSharing(final GLContext context) {
+        return new Filter<CLPlatform>() {
+            private final Filter<CLPlatform> glFilter = glSharing();
+            public boolean accept(CLPlatform item) {
+                String glVendor = context.getGL().glGetString(GL.GL_VENDOR);
+                String clVendor = item.getVendor();
+                return clVendor.equals(glVendor) && glFilter.accept(item);
+            }
+        };
+    }
+
+    /**
      * Accepts all platforms supporting the given extensions.
      */
     public static Filter<CLPlatform> extension(final String... extensions) {
@@ -98,10 +115,10 @@ public class CLPlatformFilters {
      */
     public static Filter<CLPlatform> queueMode(final Mode... modes) {
         return new Filter<CLPlatform>() {
+            private final Filter<CLDevice> queueModeFilter = CLDeviceFilters.queueMode(modes);
             public boolean accept(CLPlatform item) {
-                List<Mode> modesList = Arrays.asList(modes);
                 for (CLDevice device : item.listCLDevices()) {
-                    if(device.getQueueProperties().containsAll(modesList)) {
+                    if(queueModeFilter.accept(device)) {
                         return true;
                     }
                 }
diff --git a/src/com/jogamp/opencl/util/concurrent/CLCommandQueuePool.java b/src/com/jogamp/opencl/util/concurrent/CLCommandQueuePool.java
new file mode 100644
index 00000000..9ea960ae
--- /dev/null
+++ b/src/com/jogamp/opencl/util/concurrent/CLCommandQueuePool.java
@@ -0,0 +1,280 @@
+/*
+ * Created on Tuesday, May 03 2011
+ */
+package com.jogamp.opencl.util.concurrent;
+
+import com.jogamp.opencl.CLCommandQueue;
+import com.jogamp.opencl.CLDevice;
+import com.jogamp.opencl.CLResource;
+import com.jogamp.opencl.util.CLMultiContext;
+import java.util.ArrayList;
+import java.util.Collection;
+import java.util.List;
+import java.util.concurrent.Callable;
+import java.util.concurrent.ExecutorService;
+import java.util.concurrent.Executors;
+import java.util.concurrent.Future;
+import java.util.concurrent.ThreadFactory;
+import java.util.concurrent.TimeUnit;
+
+/**
+ * A multithreaded, fixed size pool of OpenCL command queues.
+ * It serves as a multiplexer distributing tasks over N queues usually run on N devices.
+ * The usage of this pool is similar to {@link ExecutorService} but it uses {@link CLTask}s
+ * instead of {@link Callable}s and provides a per-queue context for resource sharing across all tasks of one queue.
+ * @author Michael Bien
+ */
+public class CLCommandQueuePool<C extends CLQueueContext> implements CLResource {
+
+    private List<CLQueueContext> contexts;
+    private ExecutorService excecutor;
+    private FinishAction finishAction = FinishAction.DO_NOTHING;
+
+    private CLCommandQueuePool(CLQueueContextFactory factory, Collection<CLCommandQueue> queues) {
+        this.contexts = initContexts(queues, factory);
+        initExecutor();
+    }
+
+    private List<CLQueueContext> initContexts(Collection<CLCommandQueue> queues, CLQueueContextFactory factory) {
+        List<CLQueueContext> newContexts = new ArrayList<CLQueueContext>(queues.size());
+        
+        int index = 0;
+        for (CLCommandQueue queue : queues) {
+            
+            CLQueueContext old = null;
+            if(this.contexts != null && !this.contexts.isEmpty()) {
+                old = this.contexts.get(index++);
+                old.release();
+            }
+            
+            newContexts.add(factory.setup(queue, old));
+        }
+        return newContexts;
+    }
+
+    private void initExecutor() {
+        this.excecutor = Executors.newFixedThreadPool(contexts.size(), new QueueThreadFactory(contexts));
+    }
+
+    public static <C extends CLQueueContext> CLCommandQueuePool<C> create(CLQueueContextFactory<C> factory, CLMultiContext mc, CLCommandQueue.Mode... modes) {
+        return create(factory, mc.getDevices(), modes);
+    }
+
+    public static <C extends CLQueueContext> CLCommandQueuePool<C> create(CLQueueContextFactory<C> factory, Collection<CLDevice> devices, CLCommandQueue.Mode... modes) {
+        List<CLCommandQueue> queues = new ArrayList<CLCommandQueue>(devices.size());
+        for (CLDevice device : devices) {
+            queues.add(device.createCommandQueue(modes));
+        }
+        return create(factory, queues);
+    }
+
+    public static <C extends CLQueueContext> CLCommandQueuePool create(CLQueueContextFactory<C> factory, Collection<CLCommandQueue> queues) {
+        return new CLCommandQueuePool(factory, queues);
+    }
+
+    /**
+     * Submits this task to the pool for execution returning its {@link Future}.
+     * @see ExecutorService#submit(java.util.concurrent.Callable)
+     */
+    public <R> Future<R> submit(CLTask<? super C, R> task) {
+        return excecutor.submit(new TaskWrapper(task, finishAction));
+    }
+
+    /**
+     * Submits all tasks to the pool for execution and returns their {@link Future}.
+     * Calls {@link #submit(com.jogamp.opencl.util.concurrent.CLTask)} for every task.
+     */
+    public <R> List<Future<R>> submitAll(Collection<? extends CLTask<? super C, R>> tasks) {
+        List<Future<R>> futures = new ArrayList<Future<R>>(tasks.size());
+        for (CLTask<? super C, R> task : tasks) {
+            futures.add(submit(task));
+        }
+        return futures;
+    }
+
+    /**
+     * Submits all tasks to the pool for immediate execution (blocking) and returns their {@link Future} holding the result.
+     * @see ExecutorService#invokeAll(java.util.Collection) 
+     */
+    public <R> List<Future<R>> invokeAll(Collection<? extends CLTask<? super C, R>> tasks) throws InterruptedException {
+        List<TaskWrapper<C, R>> wrapper = wrapTasks(tasks);
+        return excecutor.invokeAll(wrapper);
+    }
+
+    /**
+     * Submits all tasks to the pool for immediate execution (blocking) and returns their {@link Future} holding the result.
+     * @see ExecutorService#invokeAll(java.util.Collection, long, java.util.concurrent.TimeUnit)
+     */
+    public <R> List<Future<R>> invokeAll(Collection<? extends CLTask<? super C, R>> tasks, long timeout, TimeUnit unit) throws InterruptedException {
+        List<TaskWrapper<C, R>> wrapper = wrapTasks(tasks);
+        return excecutor.invokeAll(wrapper, timeout, unit);
+    }
+
+    private <R> List<TaskWrapper<C, R>> wrapTasks(Collection<? extends CLTask<? super C, R>> tasks) {
+        List<TaskWrapper<C, R>> wrapper = new ArrayList<TaskWrapper<C, R>>(tasks.size());
+        for (CLTask<? super C, R> task : tasks) {
+            if(task == null) {
+                throw new NullPointerException("at least one task was null");
+            }
+            wrapper.add(new TaskWrapper<C, R>(task, finishAction));
+        }
+        return wrapper;
+    }
+    
+    /**
+     * Switches the context of all queues - this operation can be expensive.
+     * Blocks until all tasks finish and sets up a new context for all queues.
+     * @return this
+     */
+    public <C extends CLQueueContext> CLCommandQueuePool switchContext(CLQueueContextFactory<C> factory) {
+        
+        excecutor.shutdown();
+        finishQueues(); // just to be sure
+        
+        contexts = initContexts(getQueues(), factory);
+        initExecutor();
+        return this;
+    }
+
+    /**
+     * Calls {@link CLCommandQueue#flush()} on all queues.
+     */
+    public void flushQueues() {
+        for (CLQueueContext context : contexts) {
+            context.queue.flush();
+        }
+    }
+
+    /**
+     * Calls {@link CLCommandQueue#finish()} on all queues.
+     */
+    public void finishQueues() {
+        for (CLQueueContext context : contexts) {
+            context.queue.finish();
+        }
+    }
+
+    /**
+     * Releases all queues.
+     */
+    public void release() {
+        excecutor.shutdown();
+        for (CLQueueContext context : contexts) {
+            context.queue.finish().release();
+            context.release();
+        }
+    }
+
+    /**
+     * Returns the command queues used in this pool.
+     */
+    public List<CLCommandQueue> getQueues() {
+        List<CLCommandQueue> queues = new ArrayList<CLCommandQueue>(contexts.size());
+        for (CLQueueContext context : contexts) {
+            queues.add(context.queue);
+        }
+        return queues;
+    }
+
+    /**
+     * Returns the size of this pool (number of command queues).
+     */
+    public int getSize() {
+        return contexts.size();
+    }
+
+    public FinishAction getFinishAction() {
+        return finishAction;
+    }
+
+    /**
+     * Sets the action which is run after every completed task.
+     * This is mainly intended for debugging, default value is {@link FinishAction#DO_NOTHING}.
+     */
+    public void setFinishAction(FinishAction action) {
+        this.finishAction = action;
+    }
+
+    @Override
+    public String toString() {
+        return getClass().getSimpleName()+" [queues: "+contexts.size()+" on finish: "+finishAction+"]";
+    }
+
+    private static class QueueThreadFactory implements ThreadFactory {
+
+        private final List<CLQueueContext> context;
+        private int index;
+
+        private QueueThreadFactory(List<CLQueueContext> queues) {
+            this.context = queues;
+            this.index = 0;
+        }
+
+        public synchronized Thread newThread(Runnable runnable) {
+
+            SecurityManager sm = System.getSecurityManager();
+            ThreadGroup group = (sm != null) ? sm.getThreadGroup() : Thread.currentThread().getThreadGroup();
+
+            CLQueueContext queue = context.get(index);
+            QueueThread thread = new QueueThread(group, runnable, queue, index++);
+            thread.setDaemon(true);
+            
+            return thread;
+        }
+
+    }
+    
+    private static class QueueThread extends Thread {
+        private final CLQueueContext context;
+        public QueueThread(ThreadGroup group, Runnable runnable, CLQueueContext context, int index) {
+            super(group, runnable, "queue-worker-thread-"+index+"["+context+"]");
+            this.context = context;
+        }
+    }
+
+    private static class TaskWrapper<C extends CLQueueContext, R> implements Callable<R> {
+
+        private final CLTask<? super C, R> task;
+        private final FinishAction mode;
+        
+        public TaskWrapper(CLTask<? super C, R> task, FinishAction mode) {
+            this.task = task;
+            this.mode = mode;
+        }
+
+        public R call() throws Exception {
+            CLQueueContext context = ((QueueThread)Thread.currentThread()).context;
+            R result = task.execute((C)context);
+            if(mode.equals(FinishAction.FLUSH)) {
+                context.queue.flush();
+            }else if(mode.equals(FinishAction.FINISH)) {
+                context.queue.finish();
+            }
+            return result;
+        }
+
+    }
+
+    /**
+     * The action executed after a task completes.
+     */
+    public enum FinishAction {
+
+        /**
+         * Does nothing, the task is responsible to make sure all computations
+         * have finished when the task finishes
+         */
+        DO_NOTHING,
+
+        /**
+         * Flushes the queue on task completion.
+         */
+        FLUSH,
+        
+        /**
+         * Finishes the queue on task completion.
+         */
+        FINISH
+    }
+
+}
diff --git a/src/com/jogamp/opencl/util/concurrent/CLQueueContext.java b/src/com/jogamp/opencl/util/concurrent/CLQueueContext.java
new file mode 100644
index 00000000..3f89ad0e
--- /dev/null
+++ b/src/com/jogamp/opencl/util/concurrent/CLQueueContext.java
@@ -0,0 +1,72 @@
+/*
+ * Created on Friday, May 06 2011 21:02
+ */
+package com.jogamp.opencl.util.concurrent;
+
+import com.jogamp.opencl.CLCommandQueue;
+import com.jogamp.opencl.CLContext;
+import com.jogamp.opencl.CLKernel;
+import com.jogamp.opencl.CLProgram;
+import com.jogamp.opencl.CLResource;
+import java.util.Map;
+
+/**
+ * Superclass for all per-queue contexts as used in {@link CLCommandQueuePool}s.
+ * A context will usually hold queue (and therefore often device) specific resources used
+ * in tasks of the same queue.
+ * <p>
+ * Possible candidates for those resources can be compiled CLPrograms, CLKernels
+ * or even pre allocated CLBuffers.
+ * </p>
+ * @author Michael Bien
+ */
+public abstract class CLQueueContext implements CLResource {
+
+    public final CLCommandQueue queue;
+
+    public CLQueueContext(CLCommandQueue queue) {
+        this.queue = queue;
+    }
+
+    public CLCommandQueue getQueue() {
+        return queue;
+    }
+
+    public CLContext getCLContext() {
+        return queue.getContext();
+    }
+
+    /**
+     * A simple queue context holding a precompiled program and its kernels.
+     * @author Michael Bien
+     */
+    public static class CLSimpleQueueContext extends CLQueueContext {
+
+        public final CLProgram program;
+        public final Map<String, CLKernel> kernels;
+
+        public CLSimpleQueueContext(CLCommandQueue queue, CLProgram program) {
+            super(queue);
+            this.program = program;
+            this.kernels = program.createCLKernels();
+        }
+
+        public Map<String, CLKernel> getKernels() {
+            return kernels;
+        }
+
+        public CLKernel getKernel(String name) {
+            return kernels.get(name);
+        }
+
+        public CLProgram getProgram() {
+            return program;
+        }
+
+        public void release() {
+            program.release();
+        }
+
+    }
+
+}
diff --git a/src/com/jogamp/opencl/util/concurrent/CLQueueContextFactory.java b/src/com/jogamp/opencl/util/concurrent/CLQueueContextFactory.java
new file mode 100644
index 00000000..58f389bf
--- /dev/null
+++ b/src/com/jogamp/opencl/util/concurrent/CLQueueContextFactory.java
@@ -0,0 +1,51 @@
+/*
+ * Created onSaturday, May 07 2011 00:40
+ */
+package com.jogamp.opencl.util.concurrent;
+
+import com.jogamp.opencl.CLCommandQueue;
+import com.jogamp.opencl.CLProgram;
+import com.jogamp.opencl.util.concurrent.CLQueueContext.CLSimpleQueueContext;
+
+/**
+ * Creates {@link CLQueueContext}s.
+ * @author Michael Bien
+ */
+public abstract class CLQueueContextFactory<C extends CLQueueContext> {
+
+    /**
+     * Creates a new queue context for the given queue.
+     * @param old the old context or null.
+     */
+    public abstract C setup(CLCommandQueue queue, CLQueueContext old);
+
+
+    /**
+     * Creates a simple context factory producing single program contexts.
+     * @param source sourcecode of a OpenCL program.
+     */
+    public static CLSimpleContextFactory createSimple(String source) {
+        return new CLSimpleContextFactory(source);
+    }
+
+    /**
+     * Creates {@link CLSimpleQueueContext}s containing a precompiled program.
+     * @author Michael Bien
+     */
+    public static class CLSimpleContextFactory extends CLQueueContextFactory<CLSimpleQueueContext> {
+
+        private final String source;
+
+        public CLSimpleContextFactory(String source) {
+            this.source = source;
+        }
+
+        @Override
+        public CLSimpleQueueContext setup(CLCommandQueue queue, CLQueueContext old) {
+            CLProgram program = queue.getContext().createProgram(source).build(queue.getDevice());
+            return new CLSimpleQueueContext(queue, program);
+        }
+
+    }
+
+}
diff --git a/src/com/jogamp/opencl/util/concurrent/CLTask.java b/src/com/jogamp/opencl/util/concurrent/CLTask.java
new file mode 100644
index 00000000..0cfd24a5
--- /dev/null
+++ b/src/com/jogamp/opencl/util/concurrent/CLTask.java
@@ -0,0 +1,18 @@
+/*
+ * Created on Tuesday, May 03 2011 18:09
+ */
+package com.jogamp.opencl.util.concurrent;
+
+
+/**
+ * A task executed on a command queue.
+ * @author Michael Bien
+ */
+public interface CLTask<C extends CLQueueContext, R> {
+
+    /**
+     * Runs the task on a queue and returns a result.
+     */
+    R execute(C context);
+
+}
diff --git a/test/com/jogamp/opencl/CLBufferTest.java b/test/com/jogamp/opencl/CLBufferTest.java
index 1b718277..e85f4504 100644
--- a/test/com/jogamp/opencl/CLBufferTest.java
+++ b/test/com/jogamp/opencl/CLBufferTest.java
@@ -3,14 +3,14 @@
  *
  * Redistribution and use in source and binary forms, with or without modification, are
  * permitted provided that the following conditions are met:
- * 
+ *
  *    1. Redistributions of source code must retain the above copyright notice, this list of
  *       conditions and the following disclaimer.
- * 
+ *
  *    2. Redistributions in binary form must reproduce the above copyright notice, this list
  *       of conditions and the following disclaimer in the documentation and/or other materials
  *       provided with the distribution.
- * 
+ *
  * THIS SOFTWARE IS PROVIDED BY JogAmp Community ``AS IS'' AND ANY EXPRESS OR IMPLIED
  * WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND
  * FITNESS FOR A PARTICULAR PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL JogAmp Community OR
@@ -20,7 +20,7 @@
  * ANY THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
  * NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF
  * ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
- * 
+ *
  * The views and conclusions contained in the software and documentation are those of the
  * authors and should not be interpreted as representing official policies, either expressed
  * or implied, of JogAmp Community.
@@ -31,8 +31,14 @@ package com.jogamp.opencl;
 import com.jogamp.opencl.CLMemory.Mem;
 import com.jogamp.opencl.CLMemory.Map;
 import com.jogamp.common.nio.Buffers;
+import java.nio.Buffer;
 import java.nio.ByteBuffer;
+import java.nio.DoubleBuffer;
 import java.nio.FloatBuffer;
+import java.nio.IntBuffer;
+import java.nio.LongBuffer;
+import java.nio.ShortBuffer;
+import java.util.List;
 import java.util.concurrent.CountDownLatch;
 import java.util.concurrent.TimeUnit;
 import org.junit.Test;
@@ -41,6 +47,8 @@ import static org.junit.Assert.*;
 import static java.lang.System.*;
 import static com.jogamp.opencl.TestUtils.*;
 import static com.jogamp.common.nio.Buffers.*;
+import static com.jogamp.opencl.util.CLPlatformFilters.*;
+import static com.jogamp.opencl.CLVersion.*;
 
 /**
  *
@@ -48,6 +56,62 @@ import static com.jogamp.common.nio.Buffers.*;
  */
 public class CLBufferTest {
 
+
+    @Test
+    public void createBufferTest() {
+
+        out.println(" - - - highLevelTest; create buffer test - - - ");
+
+        CLContext context = CLContext.create();
+        try{
+            int size = 6;
+
+            CLBuffer<ByteBuffer> bb = context.createByteBuffer(size);
+            CLBuffer<ShortBuffer> sb = context.createShortBuffer(size);
+            CLBuffer<IntBuffer> ib = context.createIntBuffer(size);
+            CLBuffer<LongBuffer> lb = context.createLongBuffer(size);
+            CLBuffer<FloatBuffer> fb = context.createFloatBuffer(size);
+            CLBuffer<DoubleBuffer> db = context.createDoubleBuffer(size);
+
+            List<CLMemory<? extends Buffer>> buffers = context.getMemoryObjects();
+            assertEquals(6, buffers.size());
+
+            assertEquals(1, bb.getElementSize());
+            assertEquals(2, sb.getElementSize());
+            assertEquals(4, ib.getElementSize());
+            assertEquals(8, lb.getElementSize());
+            assertEquals(4, fb.getElementSize());
+            assertEquals(8, db.getElementSize());
+
+            ByteBuffer anotherNIO = newDirectByteBuffer(2);
+
+            for (CLMemory<? extends Buffer> memory : buffers) {
+
+                CLBuffer<? extends Buffer> buffer = (CLBuffer<? extends Buffer>) memory;
+                Buffer nio = buffer.getBuffer();
+
+                assertEquals(nio.capacity(), buffer.getCLCapacity());
+                assertEquals(buffer.getNIOSize(), buffer.getCLSize());
+                assertEquals(sizeOfBufferElem(nio), buffer.getElementSize());
+                assertEquals(nio.capacity() * sizeOfBufferElem(nio), buffer.getCLSize());
+                
+                CLBuffer<ByteBuffer> clone = buffer.cloneWith(anotherNIO);
+
+                assertEquals(buffer.ID, clone.ID);
+                assertTrue(clone.equals(buffer));
+                assertTrue(buffer.equals(clone));
+
+                assertEquals(buffer.getCLSize(), clone.getCLCapacity());
+                assertEquals(buffer.getCLSize(), clone.getCLSize());
+                assertEquals(anotherNIO.capacity(), clone.getNIOCapacity());
+            }
+
+        }finally{
+            context.release();
+        }
+
+    }
+
     @Test
     public void writeCopyReadBufferTest() {
 
@@ -123,7 +187,7 @@ public class CLBufferTest {
 
         context.release();
     }
-    
+
     @Test
     public void mapBufferTest() {
 
@@ -153,9 +217,9 @@ public class CLBufferTest {
         }
 
         CLCommandQueue queue = context.getDevices()[0].createCommandQueue();
-        
+
         // fill only first buffer -> we will copy the payload to the second later.
-        ByteBuffer mappedBufferA = queue.putMapBuffer(clBufferA, Map.READ_WRITE, true);
+        ByteBuffer mappedBufferA = queue.putMapBuffer(clBufferA, Map.WRITE, true);
         assertEquals(sizeInBytes, mappedBufferA.capacity());
 
         fillBuffer(mappedBufferA, 12345);                // write to A
@@ -182,21 +246,13 @@ public class CLBufferTest {
 
         out.println(" - - - subBufferTest - - - ");
 
-        CLPlatform[] platforms = CLPlatform.listCLPlatforms();
-        CLPlatform theChosenOne = null;
-        for (CLPlatform platform : platforms) {
-            if(platform.isAtLeast(CLVersion.CL_1_1)) {
-                theChosenOne = platform;
-                break;
-            }
-        }
-
-        if(theChosenOne == null) {
+        CLPlatform platform = CLPlatform.getDefault(version(CL_1_1));
+        if(platform == null) {
             out.println("aborting subBufferTest");
             return;
         }
 
-        CLContext context = CLContext.create(theChosenOne);
+        CLContext context = CLContext.create(platform);
         try{
             final int subelements = 5;
             // device only
@@ -253,8 +309,8 @@ public class CLBufferTest {
 
         out.println(" - - - destructorCallbackTest - - - ");
 
-        CLPlatform platform = CLPlatform.getDefault();
-        if(!platform.isAtLeast(CLVersion.CL_1_1)) {
+        CLPlatform platform = CLPlatform.getDefault(version(CL_1_1));
+        if(platform == null) {
             out.println("aborting destructorCallbackTest");
             return;
         }
diff --git a/test/com/jogamp/opencl/CLProgramTest.java b/test/com/jogamp/opencl/CLProgramTest.java
index 49c25180..a3ee26f3 100644
--- a/test/com/jogamp/opencl/CLProgramTest.java
+++ b/test/com/jogamp/opencl/CLProgramTest.java
@@ -268,7 +268,7 @@ public class CLProgramTest {
     @Test
     public void kernelTest() {
 
-        String source = "__attribute__((reqd_work_group_size(1, 1, 1))) kernel void foo(void) { }\n";
+        String source = "__attribute__((reqd_work_group_size(1, 1, 1))) kernel void foo(float a, int b) { }\n";
 
         CLContext context = CLContext.create();
 
@@ -287,6 +287,23 @@ public class CLProgramTest {
             assertEquals(1, wgs[1]);
             assertEquals(1, wgs[2]);
 
+            // put args test
+            assertEquals(0, kernel.position());
+
+            kernel.putArg(1.0f);
+            assertEquals(1, kernel.position());
+
+            kernel.putArg(2);
+            assertEquals(2, kernel.position());
+            
+            try{
+                kernel.putArg(3);
+                fail("exception not thrown");
+            }catch (IndexOutOfBoundsException expected){ }
+
+            assertEquals(2, kernel.position());
+            assertEquals(0, kernel.rewind().position());
+
         }finally{
             context.release();
         }
diff --git a/test/com/jogamp/opencl/LowLevelBindingTest.java b/test/com/jogamp/opencl/LowLevelBindingTest.java
index d5ef3f50..9e603879 100644
--- a/test/com/jogamp/opencl/LowLevelBindingTest.java
+++ b/test/com/jogamp/opencl/LowLevelBindingTest.java
@@ -30,7 +30,7 @@ package com.jogamp.opencl;
 
 import java.util.Random;
 import com.jogamp.opencl.impl.BuildProgramCallback;
-import com.jogamp.common.nio.PointerBuffer;
+import com.jogamp.common.nio.NativeSizeBuffer;
 
 import java.nio.ByteBuffer;
 import java.nio.IntBuffer;
@@ -109,12 +109,12 @@ public class LowLevelBindingTest {
         checkForError(ret);
         out.println("#platforms: "+intBuffer.get(0));
 
-        PointerBuffer platformId = PointerBuffer.allocateDirect(intBuffer.get(0));
+        NativeSizeBuffer platformId = NativeSizeBuffer.allocateDirect(intBuffer.get(0));
         ret = cl.clGetPlatformIDs(platformId.capacity(), platformId, null);
         checkForError(ret);
 
         // print platform info
-        PointerBuffer longBuffer = PointerBuffer.allocateDirect(1);
+        NativeSizeBuffer longBuffer = NativeSizeBuffer.allocateDirect(1);
         ByteBuffer bb = newDirectByteBuffer(128);
 
         for (int i = 0; i < platformId.capacity(); i++)  {
@@ -143,7 +143,7 @@ public class LowLevelBindingTest {
             checkForError(ret);
             out.println("#devices: "+intBuffer.get(0));
 
-            PointerBuffer devices = PointerBuffer.allocateDirect(intBuffer.get(0));
+            NativeSizeBuffer devices = NativeSizeBuffer.allocateDirect(intBuffer.get(0));
             ret = cl.clGetDeviceIDs(platform, CL.CL_DEVICE_TYPE_ALL, devices.capacity(), devices, null);
 
             //print device info
@@ -176,7 +176,7 @@ public class LowLevelBindingTest {
         checkForError(ret);
         out.println("#platforms: "+intBuffer.get(0));
 
-        PointerBuffer pb = PointerBuffer.allocateDirect(intBuffer.get(0));
+        NativeSizeBuffer pb = NativeSizeBuffer.allocateDirect(intBuffer.get(0));
         ret = cl.clGetPlatformIDs(pb.capacity(), pb, null);
         checkForError(ret);
 
@@ -187,14 +187,14 @@ public class LowLevelBindingTest {
         checkForError(ret);
         out.println("#devices: "+intBuffer.get(0));
 
-        PointerBuffer devices = PointerBuffer.allocateDirect(intBuffer.get(0));
+        NativeSizeBuffer devices = NativeSizeBuffer.allocateDirect(intBuffer.get(0));
         ret = cl.clGetDeviceIDs(platform, CL.CL_DEVICE_TYPE_ALL, devices.capacity(), devices, null);
 
         long context = cl.clCreateContext(null, devices, null, intBuffer);
         checkError("on clCreateContext", intBuffer.get());
 
         //get number of devices
-        PointerBuffer longBuffer = PointerBuffer.allocateDirect(1);
+        NativeSizeBuffer longBuffer = NativeSizeBuffer.allocateDirect(1);
         ret = cl.clGetContextInfo(context, CL.CL_CONTEXT_DEVICES, 0, null, longBuffer);
         checkError("on clGetContextInfo", ret);
 
@@ -223,19 +223,19 @@ public class LowLevelBindingTest {
         checkForError(ret);
         assertTrue(intBuffer.get(0) > 0);
 
-        PointerBuffer pb = PointerBuffer.allocateDirect(intBuffer.get(0));
+        NativeSizeBuffer pb = NativeSizeBuffer.allocateDirect(intBuffer.get(0));
         ret = cl.clGetPlatformIDs(pb.capacity(), pb, null);
         checkForError(ret);
 
         long platform = pb.get(0);
-        PointerBuffer properties = PointerBuffer.allocateDirect(3).put(CL.CL_CONTEXT_PLATFORM)
+        NativeSizeBuffer properties = NativeSizeBuffer.allocateDirect(3).put(CL.CL_CONTEXT_PLATFORM)
                                               .put(platform).put(0) // 0 terminated array
                                               .rewind();
         long context = cl.clCreateContextFromType(properties, CL.CL_DEVICE_TYPE_ALL, null, null);
         out.println("context handle: "+context);
         checkError("on clCreateContextFromType", ret);
 
-        PointerBuffer longBuffer = PointerBuffer.allocateDirect(1);
+        NativeSizeBuffer longBuffer = NativeSizeBuffer.allocateDirect(1);
         ret = cl.clGetContextInfo(context, CL.CL_CONTEXT_DEVICES, 0, null, longBuffer);
         checkError("on clGetContextInfo", ret);
 
@@ -284,7 +284,7 @@ public class LowLevelBindingTest {
 
 
         // Create the program
-        PointerBuffer lengths = PointerBuffer.allocateDirect(1).put(programSource.length()).rewind();
+        NativeSizeBuffer lengths = NativeSizeBuffer.allocateDirect(1).put(programSource.length()).rewind();
         final long program = cl.clCreateProgramWithSource(context, 1, new String[] {programSource}, lengths, intBuffer);
         out.println("program id: "+program);
         checkError("on clCreateProgramWithSource", intBuffer.get(0));
@@ -370,8 +370,8 @@ public class LowLevelBindingTest {
         checkError("on clEnqueueWriteBuffer", ret);
 
         // Launch kernel
-        PointerBuffer gWS = PointerBuffer.allocateDirect(1).put(globalWorkSize).rewind();
-        PointerBuffer lWS = PointerBuffer.allocateDirect(1).put(localWorkSize).rewind();
+        NativeSizeBuffer gWS = NativeSizeBuffer.allocateDirect(1).put(globalWorkSize).rewind();
+        NativeSizeBuffer lWS = NativeSizeBuffer.allocateDirect(1).put(localWorkSize).rewind();
         ret = cl.clEnqueueNDRangeKernel(commandQueue, kernel, 1, null, gWS, lWS, 0, null, null);
         checkError("on clEnqueueNDRangeKernel", ret);
 
diff --git a/test/com/jogamp/opencl/gl/CLGLTest.java b/test/com/jogamp/opencl/gl/CLGLTest.java
index 01fd4679..464f1468 100644
--- a/test/com/jogamp/opencl/gl/CLGLTest.java
+++ b/test/com/jogamp/opencl/gl/CLGLTest.java
@@ -32,10 +32,10 @@
 
 package com.jogamp.opencl.gl;
 
-import com.jogamp.common.os.Platform;
-import org.junit.Rule;
-import org.junit.rules.MethodRule;
-import org.junit.rules.Timeout;
+import com.jogamp.common.nio.Buffers;
+import com.jogamp.opencl.CLCommandQueue;
+import javax.media.opengl.GL2;
+import javax.media.opengl.GLException;
 import com.jogamp.opencl.CLDevice;
 import com.jogamp.newt.Display;
 import com.jogamp.newt.NewtFactory;
@@ -43,12 +43,20 @@ import com.jogamp.newt.Screen;
 import com.jogamp.newt.Window;
 import com.jogamp.newt.opengl.GLWindow;
 import com.jogamp.opencl.CLContext;
+import com.jogamp.opencl.CLMemory.Mem;
 import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.util.CLDeviceFilters;
+import com.jogamp.opencl.util.CLPlatformFilters;
+import java.nio.IntBuffer;
 import javax.media.opengl.GLCapabilities;
 import javax.media.opengl.GLProfile;
 import javax.media.opengl.GLContext;
+import org.junit.Rule;
 import org.junit.Test;
+import org.junit.rules.MethodRule;
+import org.junit.rules.Timeout;
 
+import static com.jogamp.opencl.util.CLPlatformFilters.*;
 import static org.junit.Assert.*;
 import static java.lang.System.*;
 
@@ -65,15 +73,9 @@ public class CLGLTest {
     private static GLWindow glWindow;
     private static Window window;
 
-//    @BeforeClass
-    public static void init() {
+    public static void initGL() {
         GLProfile.initSingleton(true);
 
-        // FIXME remove when JOCL is stabelized on mac
-        if(Platform.getOS().toLowerCase().contains("mac")) {
-            fail("quick exit to prevent deadlock");
-        }
-
         Display display = NewtFactory.createDisplay(null); // local display
         assertNotNull(display);
 
@@ -92,25 +94,30 @@ public class CLGLTest {
 
         glcontext = glWindow.getContext();
 //        glcontext.makeCurrent();
-        out.println("useing glcontext:");
-        out.println(glcontext);
+//        out.println(" - - - - glcontext - - - - ");
+//        out.println(glcontext);
+//        out.println(" - - - - - - - - - - - - - ");
+    }
+
+    private void deinitGL() throws GLException {
+        glcontext.release();
+        glWindow.destroy();
+        window.destroy();
+        
+        glcontext = null;
+        glWindow = null;
+        window = null;
     }
 
     @Test
     public void createContextTest() {
 
-        init();
+        initGL();
 
         out.println(" - - - glcl; createContextTest - - - ");
 
-        CLDevice[] devices = CLPlatform.getDefault().listCLDevices();
-        CLDevice device = null;
-        for (CLDevice d : devices) {
-            if(d.isGLMemorySharingSupported()) {
-                device = d;
-                break;
-            }
-        }
+        CLPlatform platform = CLPlatform.getDefault(CLPlatformFilters.glSharing());
+        CLDevice device = platform.getMaxFlopsDevice(CLDeviceFilters.glSharing());
 
         if(device == null) {
             out.println("Aborting test: no GLCL capable devices found.");
@@ -122,7 +129,7 @@ public class CLGLTest {
         out.println(device.getPlatform());
         
         assertNotNull(glcontext);
-        glcontext.makeCurrent();
+        makeGLCurrent();
         assertTrue(glcontext.isCurrent());
         
         CLContext context = CLGLContext.create(glcontext, device);
@@ -139,10 +146,96 @@ public class CLGLTest {
             // destroy cl context, gl context still current
             context.release();
             
-            glcontext.release();
-            glWindow.destroy();
+            deinitGL();
+        }
+
+    }
+    
+    @Test
+    public void vboSharing() {
+        
+        out.println(" - - - glcl; vboSharing - - - ");
+        
+        initGL();
+        makeGLCurrent();
+        assertTrue(glcontext.isCurrent());
+        
+        CLPlatform platform = CLPlatform.getDefault(glSharing(glcontext));
+        if(platform == null) {
+            out.println("test aborted");
+            return;
+        }
+        
+        CLDevice theChosenOne = platform.getMaxFlopsDevice(CLDeviceFilters.glSharing());
+        out.println(theChosenOne);
+        
+        CLGLContext context = CLGLContext.create(glcontext, theChosenOne);
+        
+        try{
+            out.println(context);
+            
+            GL2 gl = glcontext.getGL().getGL2();
+            
+            int[] id = new int[1];
+            gl.glGenBuffers(id.length, id, 0);
+            
+            IntBuffer glData = Buffers.newDirectIntBuffer(new int[] {0,1,2,3,4,5,6,7,8});
+            glData.rewind();
+
+            // create and write GL buffer
+            gl.glEnableClientState(GL2.GL_VERTEX_ARRAY);
+                gl.glBindBuffer(GL2.GL_ARRAY_BUFFER, id[0]);
+                gl.glBufferData(GL2.GL_ARRAY_BUFFER, glData.capacity()*4, glData, GL2.GL_STATIC_DRAW);
+                gl.glBindBuffer(GL2.GL_ARRAY_BUFFER, 0);
+            gl.glDisableClientState(GL2.GL_VERTEX_ARRAY);
+            gl.glFinish();
+            
+
+            // create CLGL buffer
+            IntBuffer clData = Buffers.newDirectIntBuffer(9);
+            CLGLBuffer<IntBuffer> clBuffer = context.createFromGLBuffer(clData, id[0], glData.capacity()*4, Mem.READ_ONLY);
+            
+            assertEquals(glData.capacity(), clBuffer.getCLCapacity());
+            assertEquals(glData.capacity()*4, clBuffer.getCLSize());
+                
+            
+            CLCommandQueue queue = theChosenOne.createCommandQueue();
+
+            // read gl buffer into cl nio buffer
+            queue.putAcquireGLObject(clBuffer)
+                 .putReadBuffer(clBuffer, true)
+                 .putReleaseGLObject(clBuffer);
+
+            while(clData.hasRemaining()) {
+                assertEquals(glData.get(), clData.get());
+            }
+
+            out.println(clBuffer);
+
+            clBuffer.release();
+            
+            gl.glDeleteBuffers(1, id, 0);
+            
+        }finally{
+            context.release();
+            deinitGL();
         }
+        
+    }
 
+    private void makeGLCurrent() {
+        // we are patient...
+        while(true) {
+            try{
+                glcontext.makeCurrent();
+                break;
+            }catch(RuntimeException ex) {
+                try {
+                    Thread.sleep(200);
+                    // I don't give up yet!
+                } catch (InterruptedException ignore) { }
+            }
+        }
     }
 
 
diff --git a/test/com/jogamp/opencl/util/concurrent/CLMultiContextTest.java b/test/com/jogamp/opencl/util/concurrent/CLMultiContextTest.java
new file mode 100644
index 00000000..81d34907
--- /dev/null
+++ b/test/com/jogamp/opencl/util/concurrent/CLMultiContextTest.java
@@ -0,0 +1,172 @@
+/*
+ * Created on Tuesday, May 03 2011
+ */
+package com.jogamp.opencl.util.concurrent;
+
+import com.jogamp.common.nio.Buffers;
+import com.jogamp.opencl.CLBuffer;
+import com.jogamp.opencl.CLCommandQueue;
+import com.jogamp.opencl.CLContext;
+import com.jogamp.opencl.CLDevice;
+import com.jogamp.opencl.CLKernel;
+import com.jogamp.opencl.CLPlatform;
+import com.jogamp.opencl.util.concurrent.CLQueueContext.CLSimpleQueueContext;
+import com.jogamp.opencl.util.concurrent.CLQueueContextFactory.CLSimpleContextFactory;
+import java.nio.IntBuffer;
+import java.util.concurrent.ExecutionException;
+import java.util.concurrent.Future;
+import org.junit.Rule;
+import org.junit.rules.MethodRule;
+import org.junit.rules.Timeout;
+import com.jogamp.opencl.util.CLMultiContext;
+import java.nio.Buffer;
+import java.util.ArrayList;
+import java.util.List;
+import org.junit.Test;
+
+import static org.junit.Assert.*;
+import static java.lang.System.*;
+
+/**
+ *
+ * @author Michael Bien
+ */
+public class CLMultiContextTest {
+
+    @Rule
+    public MethodRule methodTimeout= new Timeout(10000);
+
+    @Test
+    public void createMultiContextTest() {
+
+        CLMultiContext mc = CLMultiContext.create(CLPlatform.listCLPlatforms());
+
+        try{
+            List<CLContext> contexts = mc.getContexts();
+            List<CLDevice> devices = mc.getDevices();
+
+            assertFalse(contexts.isEmpty());
+            assertFalse(devices.isEmpty());
+
+            for (CLContext context : contexts) {
+                out.println(context);
+            }
+            for (CLDevice device : devices) {
+                out.println(device);
+            }
+
+        }finally{
+            mc.release();
+        }
+
+    }
+
+    private final static String programSource =
+          "kernel void compute(global int* array, int numElements) { \n"
+        + "    int index = get_global_id(0);                         \n"
+        + "    if (index >= numElements)  {                          \n"
+        + "        return;                                           \n"
+        + "    }                                                     \n"
+        + "    array[index]++;                                       \n"
+        + "}                                                         \n";
+
+    private final class CLTestTask implements CLTask<CLSimpleQueueContext, Buffer> {
+
+        private final Buffer data;
+
+        public CLTestTask(Buffer buffer) {
+            this.data = buffer;
+        }
+
+        public Buffer execute(CLSimpleQueueContext qc) {
+            
+            CLCommandQueue queue = qc.getQueue();
+            CLContext context = qc.getCLContext();
+            CLKernel kernel = qc.getKernel("compute");
+
+            CLBuffer<Buffer> buffer = null;
+            try{
+                buffer = context.createBuffer(data);
+                int gws = buffer.getCLCapacity();
+
+                kernel.putArg(buffer).putArg(gws).rewind();
+
+                queue.putWriteBuffer(buffer, true);
+                queue.put1DRangeKernel(kernel, 0, gws, 0);
+                queue.putReadBuffer(buffer, true);
+            }finally{
+                if(buffer != null) {
+                    buffer.release();
+                }
+            }
+
+            return data;
+        }
+
+    }
+
+    @Test
+    public void commandQueuePoolTest() throws InterruptedException, ExecutionException {
+
+        CLMultiContext mc = CLMultiContext.create(CLPlatform.listCLPlatforms());
+
+        try {
+
+            CLSimpleContextFactory factory = CLQueueContextFactory.createSimple(programSource);
+            CLCommandQueuePool<CLSimpleQueueContext> pool = CLCommandQueuePool.create(factory, mc);
+
+            assertTrue(pool.getSize() > 0);
+
+            final int slice = 64;
+            final int tasksPerQueue = 10;
+            final int taskCount = pool.getSize() * tasksPerQueue;
+            
+            IntBuffer data = Buffers.newDirectIntBuffer(slice*taskCount);
+
+            List<CLTestTask> tasks = new ArrayList<CLTestTask>(taskCount);
+
+            for (int i = 0; i < taskCount; i++) {
+                IntBuffer subBuffer = Buffers.slice(data, i*slice, slice);
+                assertEquals(slice, subBuffer.capacity());
+                tasks.add(new CLTestTask(subBuffer));
+            }
+
+            out.println("invoking "+tasks.size()+" tasks on "+pool.getSize()+" queues");
+
+            // blocking invoke
+            pool.invokeAll(tasks);
+            checkBuffer(1, data);
+
+            // submit blocking emediatly
+            for (CLTestTask task : tasks) {
+                pool.submit(task).get();
+            }
+            checkBuffer(2, data);
+
+            // submitAll using futures
+            List<Future<Buffer>> futures = pool.submitAll(tasks);
+            for (Future<Buffer> future : futures) {
+                future.get();
+            }
+            checkBuffer(3, data);
+
+            // switching contexts using different program
+            factory = CLQueueContextFactory.createSimple(programSource.replaceAll("\\+\\+", "--"));
+            pool.switchContext(factory);
+            pool.invokeAll(tasks);
+            checkBuffer(2, data);
+
+            pool.release();
+        }finally{
+            mc.release();
+        }
+    }
+
+    private void checkBuffer(int expected, IntBuffer data) {
+        while(data.hasRemaining()) {
+            assertEquals(expected, data.get());
+        }
+        data.rewind();
+    }
+
+}