Skip to content

Commit

Permalink
[+] Multiple .cl files can be merged into one by linking them together
Browse files Browse the repository at this point in the history
[*] Changed naming convention
[!] clLinkProgram crashes JVM on Windows: test further (see LWJGL/lwjgl3#560)
  • Loading branch information
mikailgedik committed Dec 16, 2020
1 parent 6cd0f5d commit 498ae35
Show file tree
Hide file tree
Showing 10 changed files with 314 additions and 86 deletions.
74 changes: 21 additions & 53 deletions src/main/java/com/rayx/RayX.java
Original file line number Diff line number Diff line change
Expand Up @@ -13,7 +13,6 @@ public class RayX {
static CLContext context;

public static void main(String[] args) {

WindowManager manager = WindowManager.getInstance();

TestOGLWindow window = new TestOGLWindow(800, 600, "Test");
Expand All @@ -38,7 +37,7 @@ public static void main(String[] args) {

window.setCallback((texture) -> {
glFinish();
CLContext.CLKernel kernel = context.getKernelObject("test");
CLContext.CLKernel kernel = context.getKernelObject("testKernel");
CLManager.allocateMemory(context, CL_MEM_READ_ONLY, new double[]{-1.5, 0.0, .5}, "center");
CLManager.createCLFromGLTexture(context, CL_MEM_WRITE_ONLY, GL_TEXTURE_2D, texture, "tex");
context.getMemoryObject("tex").acquireFromGL();
Expand All @@ -55,71 +54,40 @@ public static void main(String[] args) {
freeAll();
}

static final String KERNEL_SOURCE = """
float4 color(int, int);
int calc(double2, int);
__kernel void testKernel(__write_only image2d_t image, __global double* center) {
int2 pixCo = (int2){get_global_id(0),get_global_id(1)};
int w = get_image_width(image);
int h = get_image_height(image);
if(pixCo.x >= w | pixCo.y >= h) {
return;
}
double2 absCo = (double2)
{center[2] * ((1.0 * pixCo.x / w) - .5) + center[0],
center[2] * ((1.0 * (h - pixCo.y - 1) / h) - .5) + center[1]};
write_imagef(image, pixCo, color(calc(absCo, 100000), 100000));
}
int calc(double2 coo, int maxIter) {
double a = 0, b = 0, ta, tb;
int i = 0;
for(;(i < maxIter) & !(a * a + b * b > 4); i++) {
ta = a*a - b*b + coo.x;
tb = 2 * a * b + coo.y;
a = ta;
b = tb;
}
return i == maxIter ? -1 : i;
}
float4 color(int value, int maxIterations) {
int color;
if(value == -1) {
color = 0x0;
} else if(value == 0) {
color = 0x0;
} else {
color = 0xffffff * (1.0*log((double)value)/log((double)maxIterations));
}
return (float4){
1.0 * ((color >> 0) & 0xFF) / 0xFF,
1.0 * ((color >> 8) & 0xFF) / 0xFF,
1.0 * ((color >> 16) & 0xFF) / 0xFF,
1
};
}
""";

static CLContext treatDevice(OpenGLWindow window, long device) {
long t = System.currentTimeMillis();
CLContext context = CLManager.createContext(device, window.getWindow());

System.out.println("Context: " + (System.currentTimeMillis() - t));
t = System.currentTimeMillis();

CLManager.addProgramAndKernelToContext(context, KERNEL_SOURCE, "testKernel", "test");
CLManager.putProgramFromFile(context, null,
"clcode/headers/utility/mandelbrot.h");
CLManager.putProgramFromFile(context, new String[]{
"clcode/headers/utility/mandelbrot.h"
},
"clcode/utility/mandelbrot.cl");
CLManager.putProgramFromFile(context, new String[]{
"clcode/headers/utility/mandelbrot.h"
},
"clcode/main.cl");
CLManager.putExecutableProgram(context, new String[] {
"clcode/utility/mandelbrot.cl",
//Can be added, does not change anything
//"clcode/headers/utility/mandelbrot.h",
"clcode/main.cl",
},"testProgram");

CLManager.putKernel(context, "testKernel", "testKernel", "testProgram");

System.out.println("Kernel: " + (System.currentTimeMillis() - t));
return context;
}

static void freeAll() {
long t = System.currentTimeMillis();

context.destroyKernel("test");
context.destroyKernel("testKernel");
context.destroy();
System.out.println("Destroy: " + (System.currentTimeMillis() - t));
}
Expand Down
2 changes: 0 additions & 2 deletions src/main/java/com/rayx/examples/TestOGLWindow.java
Original file line number Diff line number Diff line change
Expand Up @@ -42,7 +42,6 @@ void main()
}

private int texture;
private long window;
private int VAO, EBO, VBO;
private long lastPrint = 0;
private int frames;
Expand Down Expand Up @@ -116,7 +115,6 @@ private int createShader() {
int vs = compileShader(GL_VERTEX_SHADER, VS);
int fs = compileShader(GL_FRAGMENT_SHADER, FS);


glAttachShader(program, vs);
glAttachShader(program, fs);
glLinkProgram(program);
Expand Down
94 changes: 73 additions & 21 deletions src/main/java/com/rayx/opencl/CLContext.java
Original file line number Diff line number Diff line change
Expand Up @@ -6,8 +6,11 @@
import java.util.HashMap;

public class CLContext {
private long context, device, commandQueue;
private final long device;
private long context;
private long commandQueue;
private HashMap<String, CLKernel> kernels;
private HashMap<String, CLProgram> programs;
private HashMap<String, CLMemoryObject> memoryObjects;

/** Only {@link CLManager} should create instances */
Expand All @@ -16,6 +19,7 @@ public class CLContext {
this.device = device;
this.commandQueue = commandQueue;
kernels = new HashMap<>();
programs = new HashMap<>();
memoryObjects = new HashMap<>();
}

Expand All @@ -26,14 +30,15 @@ public void freeMemoryObject(String id) {
}

public void destroy() {
for(CLKernel k: kernels.values()) {
k.destroy();
}
kernels.values().forEach(CLKernel::destroy);
kernels.clear();
kernels = null;

memoryObjects.values().forEach(CLMemoryObject::delete);
programs.values().forEach(CLProgram::destroy);
programs.clear();
programs = null;

memoryObjects.values().forEach(CLMemoryObject::delete);
memoryObjects.clear();
memoryObjects = null;

Expand All @@ -48,6 +53,11 @@ public void getMemoryObjectValue(String id, ByteBuffer destination) {
getMemoryObject(id).getValue(destination);
}

public CLProgram getProgramObject(String id) {
assert programs.containsKey(id);
return programs.get(id);
}

public CLMemoryObject getMemoryObject(String id) {
assert memoryObjects.containsKey(id);
return memoryObjects.get(id);
Expand All @@ -70,9 +80,21 @@ public long getCommandQueue() {
return commandQueue;
}

public void addKernelObject(String id, CLKernel kernel) {
assert !kernels.containsKey(id);
kernels.put(id, kernel);
public void destroyProgram(String name) {
assert kernels.values().stream().
noneMatch(k -> name.equals(k.getProgramId())) : "Kernels depend on program \"" + name + "\"";
assert kernels.containsKey(name);
kernels.remove(name).destroy();
}

public void addKernelObject(CLKernel kernel) {
assert !kernels.containsKey(kernel.kernelId);
kernels.put(kernel.kernelId, kernel);
}

public void addProgramObject(CLProgram program) {
assert !programs.containsKey(program.programId);
programs.put(program.programId, program);
}

public CLKernel getKernelObject(String id) {
Expand All @@ -91,7 +113,7 @@ public class CLMemoryObject {
//In bytes
//If size is -1, it indicates that the memory object should not be read to the main memory
//This is the case when
private long size;
private final long size;

public CLMemoryObject(long pointer, long size) {
this.pointer = pointer;
Expand Down Expand Up @@ -126,14 +148,46 @@ public long getSize() {
}
}

public class CLKernel {
private String name;
public class CLProgram {
private final String programId;
private long program;
/** A kernel can only be created from a program if <code>linked</code> is set to <code>true</code>*/
private final boolean linked;

/** @param programId Name of the program.
* Should be name of the file it was generated from,
* so that it may used for include directives in OpenCL C*/
public CLProgram(String programId, long program, boolean linked) {
this.programId = programId;
this.program = program;
this.linked = linked;
}

public String getProgramId() {
return programId;
}

public long getProgram() {
return program;
}

private void destroy() {
CLManager.destroyCLProgramInternal(program);
program = 0;
}

public boolean isLinked() {
return linked;
}
}

public class CLKernel {
private final String kernelId, programId;
private long kernel;

public CLKernel(String name, long program, long kernel) {
this.name = name;
this.program = program;
public CLKernel(String kernelId, String programId, long kernel) {
this.kernelId = kernelId;
this.programId = programId;
this.kernel = kernel;
}

Expand All @@ -155,22 +209,20 @@ public void run(long[] globalWorkSize, long[] localWorkSize) {
}

private void destroy() {
CLManager.destroyCLKernelInternal(program, kernel);
CLManager.destroyCLKernelInternal(kernel);
this.kernel = 0;
this.program = 0;
name = null;
}

public long getKernel() {
return kernel;
}

public long getProgram() {
return program;
public String getProgramId() {
return programId;
}

public String getName() {
return name;
public String getKernelId() {
return kernelId;
}
}
}
Loading

0 comments on commit 498ae35

Please sign in to comment.