resolve conflicts

26 files changed, 1098 insertions, 312 deletions

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);
+
+}