summaryrefslogtreecommitdiffstats
path: root/test
diff options
context:
space:
mode:
Diffstat (limited to 'test')
-rw-r--r--test/com/jogamp/opencl/CLBufferTest.java8
-rw-r--r--test/com/jogamp/opencl/CLCommandQueueTest.java81
-rw-r--r--test/com/jogamp/opencl/testkernels.cl23
3 files changed, 96 insertions, 16 deletions
diff --git a/test/com/jogamp/opencl/CLBufferTest.java b/test/com/jogamp/opencl/CLBufferTest.java
index 0e4a4a65..1b718277 100644
--- a/test/com/jogamp/opencl/CLBufferTest.java
+++ b/test/com/jogamp/opencl/CLBufferTest.java
@@ -158,10 +158,10 @@ public class CLBufferTest {
ByteBuffer mappedBufferA = queue.putMapBuffer(clBufferA, Map.READ_WRITE, true);
assertEquals(sizeInBytes, mappedBufferA.capacity());
- fillBuffer(mappedBufferA, 12345); // write to A
+ fillBuffer(mappedBufferA, 12345); // write to A
- queue.putUnmapMemory(clBufferA) // unmap A
- .putCopyBuffer(clBufferA, clBufferB); // copy A -> B
+ queue.putUnmapMemory(clBufferA, mappedBufferA)// unmap A
+ .putCopyBuffer(clBufferA, clBufferB); // copy A -> B
// map B for read operations
ByteBuffer mappedBufferB = queue.putMapBuffer(clBufferB, Map.READ, true);
@@ -171,7 +171,7 @@ public class CLBufferTest {
checkIfEqual(mappedBufferA, mappedBufferB, elements); // A == B ?
out.println("results are valid");
- queue.putUnmapMemory(clBufferB); // unmap B
+ queue.putUnmapMemory(clBufferB, mappedBufferB); // unmap B
context.release();
diff --git a/test/com/jogamp/opencl/CLCommandQueueTest.java b/test/com/jogamp/opencl/CLCommandQueueTest.java
index e40d07e4..c8c028af 100644
--- a/test/com/jogamp/opencl/CLCommandQueueTest.java
+++ b/test/com/jogamp/opencl/CLCommandQueueTest.java
@@ -35,8 +35,11 @@ import java.util.concurrent.CountDownLatch;
import com.jogamp.opencl.util.MultiQueueBarrier;
import com.jogamp.opencl.CLCommandQueue.Mode;
import com.jogamp.opencl.CLMemory.Mem;
+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 java.util.EnumSet;
import java.util.concurrent.TimeUnit;
import org.junit.Test;
@@ -47,6 +50,7 @@ import static com.jogamp.opencl.TestUtils.*;
import static com.jogamp.opencl.CLEvent.*;
import static com.jogamp.opencl.CLVersion.*;
import static com.jogamp.common.nio.Buffers.*;
+import static com.jogamp.opencl.CLCommandQueue.Mode.*;
/**
*
@@ -62,8 +66,8 @@ public class CLCommandQueueTest {
//CLCommandQueueEnums
EnumSet<Mode> queueMode = Mode.valuesOf(CL.CL_QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE | CL.CL_QUEUE_PROFILING_ENABLE);
- assertTrue(queueMode.contains(Mode.OUT_OF_ORDER_MODE));
- assertTrue(queueMode.contains(Mode.PROFILING_MODE));
+ assertTrue(queueMode.contains(OUT_OF_ORDER_MODE));
+ assertTrue(queueMode.contains(PROFILING_MODE));
assertNotNull(Mode.valuesOf(0));
assertEquals(0, Mode.valuesOf(0).size());
@@ -137,11 +141,15 @@ public class CLCommandQueueTest {
queue.put1DRangeKernel(vectorAddKernel, 0, elements, groupSize, events);
assertEquals(2, events.size());
- queue.putWaitForEvent(events, 0, false)
+ queue.putWaitForEvent(events, 0, true)
.putWaitForEvent(events, 1, true);
- queue.putReadBuffer(clBufferC, false)
- .putReadBuffer(clBufferD, true);
+ events.release();
+
+ queue.putReadBuffer(clBufferC, false, events)
+ .putReadBuffer(clBufferD, false, events);
+
+ queue.putWaitForEvents(events, true);
events.release();
@@ -151,6 +159,67 @@ public class CLCommandQueueTest {
context.release();
}
}
+
+ @Test
+ public void eventConditionsTest() throws IOException {
+
+ out.println(" - - - event conditions test - - - ");
+
+ CLPlatform platform = CLPlatform.getDefault(CLPlatformFilters.queueMode(OUT_OF_ORDER_MODE));
+
+ CLDevice device = null;
+ // we can still test this with in-order queues
+ if(platform == null) {
+ device = CLPlatform.getDefault().getMaxFlopsDevice();
+ }else{
+ device = platform.getMaxFlopsDevice(CLDeviceFilters.queueMode(OUT_OF_ORDER_MODE));
+ }
+
+ CLContext context = CLContext.create(device);
+
+ try{
+
+ CLProgram program = context.createProgram(getClass().getResourceAsStream("testkernels.cl")).build();
+
+ CLBuffer<IntBuffer> buffer = context.createBuffer(newDirectIntBuffer(new int[]{ 1,1,1, 1,1,1, 1,1,1 }));
+
+ int elements = buffer.getNIOCapacity();
+
+ CLCommandQueue queue;
+ if(device.getQueueProperties().contains(OUT_OF_ORDER_MODE)) {
+ queue = device.createCommandQueue(OUT_OF_ORDER_MODE);
+ }else{
+ queue = device.createCommandQueue();
+ }
+
+ CLEventList writeEvent = new CLEventList(1);
+ CLEventList kernelEvents = new CLEventList(2);
+
+ // (1+1)*2 = 4; conditions enforce propper order
+ CLKernel addKernel = program.createCLKernel("add").putArg(buffer).putArg(1).putArg(elements);
+ CLKernel mulKernel = program.createCLKernel("mul").putArg(buffer).putArg(2).putArg(elements);
+
+ queue.putWriteBuffer(buffer, false, writeEvent);
+
+ queue.put1DRangeKernel(addKernel, 0, elements, 1, writeEvent, kernelEvents);
+ queue.put1DRangeKernel(mulKernel, 0, elements, 1, writeEvent, kernelEvents);
+
+ queue.putReadBuffer(buffer, false, kernelEvents, null);
+
+ queue.finish();
+
+ writeEvent.release();
+ kernelEvents.release();
+
+ for (int i = 0; i < elements; i++) {
+ assertEquals(4, buffer.getBuffer().get(i));
+ }
+
+ }finally{
+ context.release();
+ }
+
+ }
@Test
public void profilingEventsTest() throws IOException {
@@ -174,7 +243,7 @@ public class CLCommandQueueTest {
CLProgram program = context.createProgram(getClass().getResourceAsStream("testkernels.cl")).build();
CLKernel vectorAddKernel = program.createCLKernel("VectorAddGM").setArg(3, elements);
- CLCommandQueue queue = device.createCommandQueue(Mode.PROFILING_MODE);
+ CLCommandQueue queue = device.createCommandQueue(PROFILING_MODE);
out.println(queue);
diff --git a/test/com/jogamp/opencl/testkernels.cl b/test/com/jogamp/opencl/testkernels.cl
index ec7e8bf6..2b8c097d 100644
--- a/test/com/jogamp/opencl/testkernels.cl
+++ b/test/com/jogamp/opencl/testkernels.cl
@@ -1,22 +1,33 @@
- // OpenCL Kernel Function for element by element vector addition
kernel void VectorAddGM(global const int* a, global const int* b, global int* c, int iNumElements) {
- // get index into global data array
int iGID = get_global_id(0);
- // bound check (equivalent to the limit on a 'for' loop for standard/serial C code
if (iGID >= iNumElements) {
return;
}
- // add the vector elements
c[iGID] = a[iGID] + b[iGID];
}
kernel void Test(global const int* a, global const int* b, global int* c, int iNumElements) {
- // get index into global data array
int iGID = get_global_id(0);
- // bound check (equivalent to the limit on a 'for' loop for standard/serial C code
if (iGID >= iNumElements) {
return;
}
c[iGID] = iGID;
}
+
+ kernel void add(global int* a, int value, int iNumElements) {
+ int iGID = get_global_id(0);
+ if (iGID >= iNumElements) {
+ return;
+ }
+ a[iGID] += value;
+ }
+
+ kernel void mul(global int* a, int value, int iNumElements) {
+
+ int iGID = get_global_id(0);
+ if (iGID >= iNumElements) {
+ return;
+ }
+ a[iGID] *= value;
+ }
generated by cgit v1.2.3 (git 2.39.1) at 2024-11-29 04:43:15 +0000