diff options
Diffstat (limited to 'test/com/jogamp/opencl')
| -rw-r--r-- | test/com/jogamp/opencl/LowLevelBindingTest.java | 5 | ||||
| -rw-r--r-- | test/com/jogamp/opencl/gl/CLGLTest.java | 115 |
2 files changed, 119 insertions, 1 deletions
diff --git a/test/com/jogamp/opencl/LowLevelBindingTest.java b/test/com/jogamp/opencl/LowLevelBindingTest.java index a4cbd03e..ef14534a 100644 --- a/test/com/jogamp/opencl/LowLevelBindingTest.java +++ b/test/com/jogamp/opencl/LowLevelBindingTest.java @@ -253,7 +253,7 @@ public class LowLevelBindingTest extends UITestCase { // Was originally 4096, but had to make this bigger or it would crash in UITestCase.oneTimeTearDown(){ System.gc() } // without even dumping a stack when using AMD drivers. Presumably the drivers would write past the end // of the block and mess up GC info somehow. - ByteBuffer bb = newDirectByteBuffer(8192); + ByteBuffer bb = newDirectByteBuffer(32768); ret = cl.clGetContextInfo(context, CL.CL_CONTEXT_DEVICES, bb.capacity(), bb, null); checkError("on clGetContextInfo", ret); @@ -267,6 +267,7 @@ public class LowLevelBindingTest extends UITestCase { offset *= (is32Bit() ? 4 : 8); long device = is32Bit()?bb.getInt(offset):bb.getLong(offset); + bb.rewind(); ret = cl.clGetDeviceInfo(device, CL.CL_DEVICE_MAX_WORK_GROUP_SIZE, bb.capacity(), bb, null); checkError("on clGetDeviceInfo", ret); int maxWGS = bb.getInt(); @@ -352,11 +353,13 @@ public class LowLevelBindingTest extends UITestCase { out.println("program source length (cl): "+longBuffer.get(0)); out.println("program source length (java): "+programSource.length()); + bb.rewind(); ret = cl.clGetProgramInfo(program, CL.CL_PROGRAM_SOURCE, bb.capacity(), bb, null); checkError("on clGetProgramInfo CL_PROGRAM_SOURCE", ret); out.println("program source:\n" + clString2JavaString(bb, (int)longBuffer.get(0))); // Check program status + bb.rewind(); ret = cl.clGetProgramBuildInfo(program, device, CL.CL_PROGRAM_BUILD_STATUS, bb.capacity(), bb, null); checkError("on clGetProgramBuildInfo1", ret); diff --git a/test/com/jogamp/opencl/gl/CLGLTest.java b/test/com/jogamp/opencl/gl/CLGLTest.java index 2ad31aa3..0d1a8c3f 100644 --- a/test/com/jogamp/opencl/gl/CLGLTest.java +++ b/test/com/jogamp/opencl/gl/CLGLTest.java @@ -33,7 +33,10 @@ package com.jogamp.opencl.gl; import com.jogamp.common.nio.Buffers; +import com.jogamp.opencl.CLBuffer; import com.jogamp.opencl.CLCommandQueue; +import com.jogamp.opencl.CLKernel; +import com.jogamp.opencl.CLProgram; import javax.media.opengl.GL2; import javax.media.opengl.GLException; @@ -51,6 +54,7 @@ import com.jogamp.opencl.util.CLDeviceFilters; import com.jogamp.opencl.util.CLPlatformFilters; import java.io.IOException; +import java.nio.ByteBuffer; import java.nio.IntBuffer; import javax.media.opengl.GLCapabilities; @@ -222,6 +226,117 @@ public class CLGLTest extends UITestCase { } + @Test(timeout=15000) + public void textureSharing() { + + out.println(" - - - glcl; textureSharing - - - "); + if(MiscUtils.isOpenCLUnavailable()) + return; + + initGL(); + makeGLCurrent(); + assertTrue(glcontext.isCurrent()); + + @SuppressWarnings("unchecked") + CLPlatform [] clplatforms = CLPlatform.listCLPlatforms(glSharing(glcontext)); + if(clplatforms.length == 0) { + out.println("no platform that supports OpenGL-OpenCL interoperability"); + return; + } + + for(CLPlatform clplatform : clplatforms) { + + @SuppressWarnings("unchecked") + CLDevice [] cldevices = clplatform.listCLDevices(CLDeviceFilters.glSharing()); + + for(CLDevice cldevice : cldevices) { + out.println(cldevice); + textureSharingInner(cldevice); + } + } + + deinitGL(); + } + + public void textureSharingInner(CLDevice cldevice) { + + CLGLContext clglcontext = CLGLContext.create(glcontext, cldevice); + + try { + out.println(clglcontext); + + GL2 gl = glcontext.getGL().getGL2(); + + // create and write GL texture + int[] id = new int[1]; + gl.glGenTextures(id.length, id, 0); + gl.glActiveTexture(GL2.GL_TEXTURE0); + gl.glBindTexture (GL2.GL_TEXTURE_2D, id[0]); + int texWidth = 2; + int texHeight = 2; + gl.glTexImage2D(GL2.GL_TEXTURE_2D, 0, GL2.GL_RGBA, texWidth, texHeight, 0, GL2.GL_RGBA, GL2.GL_UNSIGNED_BYTE, null ); + gl.glBindTexture(GL2.GL_TEXTURE_2D, 0); + gl.glFinish(); + + // create CLGL buffer + ByteBuffer bufferCL = Buffers.newDirectByteBuffer(texWidth*texHeight*4); + CLGLTexture2d<ByteBuffer> clTexture = clglcontext.createFromGLTexture2d(bufferCL, GL2.GL_TEXTURE_2D, id[0], 0, CLBuffer.Mem.WRITE_ONLY); + + // set texel values to a formula that can be read back and verified + String sourceCL = "__kernel void writeTexture (__write_only image2d_t imageTex, unsigned w, unsigned h ) \n" + + "{ \n" + + " for(unsigned y=1; y<=h; ++y) { \n" + + " for(unsigned x=1; x<=w; ++x) { \n" + + " write_imagef(imageTex, (int2)(x-1,y-1), (float4)(((float)x)/((float)(4*w)), ((float)y)/((float)(4*h)), 0.0f, 1.0f)); \n" + + " } \n" + + " } \n" + + "}"; + CLProgram program = clglcontext.createProgram(sourceCL); + program.build(); + System.out.println(program.getBuildStatus()); + System.out.println(program.getBuildLog()); + assertTrue(program.isExecutable()); + + CLKernel clkernel = program.createCLKernel("writeTexture") + .putArg(clTexture) + .putArg(texWidth) + .putArg(texHeight) + .rewind(); + + CLCommandQueue queue = cldevice.createCommandQueue(); + + // write gl texture with cl kernel, then read it to host buffer + queue.putAcquireGLObject(clTexture) + .put1DRangeKernel(clkernel, 0, 1, 1) + .putReadImage(clTexture, true) + .putReleaseGLObject(clTexture) + .finish(); + + for(int y = 1; y <= texHeight; y++) { + for(int x = 1; x <= texWidth; x++) { + byte bX = bufferCL.get(); + byte bY = bufferCL.get(); + byte bZero = bufferCL.get(); + byte bMinusOne = bufferCL.get(); + byte bXCheck = (byte)(((float)x)/((float)(4*texWidth))*256); + byte bYCheck = (byte)(((float)y)/((float)(4*texHeight))*256); + assertEquals(bXCheck, bX); + assertEquals(bYCheck, bY); + assertEquals(0, bZero); + assertEquals(-1, bMinusOne); + } + } + + out.println(clTexture); + + clTexture.release(); + gl.glDeleteBuffers(1, id, 0); + } + finally { + clglcontext.release(); + } + } + private void makeGLCurrent() { // we are patient... while(true) { |