studiorailgun/fluid-sim
2
0
Fork 0
Code Issues Pull Requests Packages Projects Releases Wiki Activity

JNI attempt 1

Browse Source
This commit is contained in:
unknown 2023-07-13 10:34:54 -04:00
parent 072d4f7784
commit 63e3eeba79
28 changed files with 549 additions and 1837 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

28
dependency-reduced-pom.xml
View File

@ -2,7 +2,7 @@
<project xmlns="http://maven.apache.org/POM/4.0.0" xmlns:xsi="http://www.w3.org/2001/XMLSchema-instance" xsi:schemaLocation="http://maven.apache.org/POM/4.0.0 http://maven.apache.org/maven-v4_0_0.xsd"> <project xmlns="http://maven.apache.org/POM/4.0.0" xmlns:xsi="http://www.w3.org/2001/XMLSchema-instance" xsi:schemaLocation="http://maven.apache.org/POM/4.0.0 http://maven.apache.org/maven-v4_0_0.xsd">
<modelVersion>4.0.0</modelVersion> <modelVersion>4.0.0</modelVersion>
<groupId>electrosphere</groupId> <groupId>electrosphere</groupId>
<artifactId>sample-cli-java</artifactId> <artifactId>fluid-sim</artifactId>
<version>1.0-SNAPSHOT</version> <version>1.0-SNAPSHOT</version>
<build> <build>
<plugins> <plugins>
@ -55,11 +55,31 @@
<source>17</source> <source>17</source>
<target>17</target> <target>17</target>
<compilerArgs> <compilerArgs>
<arg>--add-modules</arg> <arg>-h</arg>
<arg>jdk.incubator.vector</arg> <arg>src/main/c/includes</arg>
</compilerArgs> </compilerArgs>
</configuration> </configuration>
</plugin> </plugin>
<plugin>
<groupId>org.codehaus.mojo</groupId>
<artifactId>exec-maven-plugin</artifactId>
<version>3.1.0</version>
<executions>
<execution>
<id>compile-c-code</id>
<phase>generate-resources</phase>
<goals>
<goal>exec</goal>
</goals>
<configuration>
<executable>bash</executable>
<arguments>
<argument>./src/main/c/compile.sh</argument>
</arguments>
</configuration>
</execution>
</executions>
</plugin>
</plugins> </plugins>
</build> </build>
<profiles> <profiles>
@ -99,6 +119,6 @@
<maven.compiler.source>17</maven.compiler.source> <maven.compiler.source>17</maven.compiler.source>
<project.build.sourceEncoding>UTF-8</project.build.sourceEncoding> <project.build.sourceEncoding>UTF-8</project.build.sourceEncoding>
<joml.version>1.10.5</joml.version> <joml.version>1.10.5</joml.version>
<lwjgl.version>3.3.2</lwjgl.version> <lwjgl.version>3.2.3</lwjgl.version>
</properties> </properties>
</project> </project>

27
pom.xml
View File

@ -9,7 +9,7 @@
<project.build.sourceEncoding>UTF-8</project.build.sourceEncoding> <project.build.sourceEncoding>UTF-8</project.build.sourceEncoding>
<maven.compiler.source>17</maven.compiler.source> <maven.compiler.source>17</maven.compiler.source>
<maven.compiler.target>17</maven.compiler.target> <maven.compiler.target>17</maven.compiler.target>
<lwjgl.version>3.3.2</lwjgl.version> <lwjgl.version>3.2.3</lwjgl.version>
<joml.version>1.10.5</joml.version> <joml.version>1.10.5</joml.version>
</properties> </properties>
@ -133,7 +133,6 @@
</execution> </execution>
</executions> </executions>
</plugin> </plugin>
<plugin> <plugin>
<artifactId>maven-assembly-plugin</artifactId> <artifactId>maven-assembly-plugin</artifactId>
<version>3.0.0</version> <version>3.0.0</version>
@ -164,11 +163,31 @@
<source>17</source> <source>17</source>
<target>17</target> <target>17</target>
<compilerArgs> <compilerArgs>
<arg>--add-modules</arg> <arg>-h</arg>
<arg>jdk.incubator.vector</arg> <arg>src/main/c/includes</arg>
</compilerArgs> </compilerArgs>
</configuration> </configuration>
</plugin> </plugin>
<plugin>
<groupId>org.codehaus.mojo</groupId>
<artifactId>exec-maven-plugin</artifactId>
<version>3.1.0</version>
<executions>
<execution>
<id>compile-c-code</id>
<phase>generate-resources</phase>
<goals>
<goal>exec</goal>
</goals>
<configuration>
<executable>bash</executable>
<arguments>
<argument>./src/main/c/compile.sh</argument>
</arguments>
</configuration>
</execution>
</executions>
</plugin>
</plugins> </plugins>
</build> </build>

BIN
shared-folder/libfluidsim.dll Normal file
View File

Binary file not shown.

62
src/main/c/compile.sh Normal file
View File

@ -0,0 +1,62 @@
cd ./src/main/c
LIB_ENDING=".so"
BASE_INCLUDE_DIR=""
OS_INCLUDE_DIR=""
if [[ "$OSTYPE" == "linux-gnu"* ]]; then
#linux
LIB_ENDING=".so"
BASE_INCLUDE_DIR="${JAVA_HOME}/include"
OS_INCLUDE_DIR="${JAVA_HOME}/include/linux"
elif [[ "$OSTYPE" == "darwin"* ]]; then
# Mac OSX
LIB_ENDING=".so"
elif [[ "$OSTYPE" == "cygwin" ]]; then
# POSIX compatibility layer and Linux environment emulation for Windows
LIB_ENDING=".dll"
elif [[ "$OSTYPE" == "msys" ]]; then
# Lightweight shell and GNU utilities compiled for Windows (part of MinGW)
LIB_ENDING=".dll"
BASE_INCLUDE_DIR="${JAVA_HOME}/include"
OS_INCLUDE_DIR="${JAVA_HOME}/include/win32"
elif [[ "$OSTYPE" == "win32" ]]; then
# I'm not sure this can happen.
LIB_ENDING=".dll"
elif [[ "$OSTYPE" == "freebsd"* ]]; then
# ...
LIB_ENDING=".so"
else
# Unknown.
LIB_ENDING=".so"
fi
# echo $PWD
#clean compile dir
rm -f ./*.o
rm -f ./*.so
rm -f ./*.dll
#compile object files
COMPILE_FLAGS="-c -fPIC -m64"
INPUT_FILES="./fluidsim.c"
OUTPUT_FILE="./fluidsim.o"
gcc $COMPILE_FLAGS -I"$BASE_INCLUDE_DIR" -I"$OS_INCLUDE_DIR" $INPUT_FILES -o $OUTPUT_FILE
#compile shared object file
OUTPUT_FILE="libfluidsim$LIB_ENDING"
COMPILE_FLAGS="-shared"
INPUT_FILES="fluidsim.o"
gcc $COMPILE_FLAGS $INPUT_FILES -o $OUTPUT_FILE
#move to resources
mv "./libfluidsim$LIB_ENDING" "../../../shared-folder/"
#clean compile dir
rm -f ./*.o
rm -f ./*.so
rm -f ./*.dll

237
src/main/c/fluidsim.c Normal file
View File

@ -0,0 +1,237 @@
#include <jni.h>
#include <stdio.h>
#define SWAP(x0,x) {float *tmp=x0;x0=x;x=tmp;}
#define IX(i,j,k) ((i)+(N)*(j)+(N*N)*(k))
#define LINEARSOLVERTIMES 20
void diffuse(int N, int b, float * x, float * x0, float diff, float dt);
void advect(int N, int b, float * d, float * d0, float * u, float * v, float * w, float dt);
void project(int N, float * u, float * v, float * w, float * p, float * div);
void set_bnd(int N, int b, float * x);
void dens_step(int N, float * x, float * x0, float * u, float * v, float * w, float diff, float dt);
void vel_step(int N, float * u, float * v, float * w, float * u0, float * v0, float * w0, float visc, float dt);
void lin_solve(int N, int b, float* x, float* x0, float a, float c);
JNIEXPORT void JNICALL Java_electrosphere_FluidSim_simulate(
JNIEnv * env,
jobject this,
jint DIM_X,
jint DIM_Y,
jint DIM_Z,
jfloatArray jx,
jfloatArray jx0,
jfloatArray ju,
jfloatArray jv,
jfloatArray jw,
jfloatArray ju0,
jfloatArray jv0,
jfloatArray jw0,
jfloat DIFFUSION_RATE,
jfloat VISCOSITY_RATE,
jfloat timestep){
jboolean isCopy;
float * x = (*env)->GetFloatArrayElements(env,jx,&isCopy);
float * x0 = (*env)->GetFloatArrayElements(env,jx0,&isCopy);
float * u = (*env)->GetFloatArrayElements(env,ju,&isCopy);
float * v = (*env)->GetFloatArrayElements(env,jv,&isCopy);
float * w = (*env)->GetFloatArrayElements(env,jw,&isCopy);
float * u0 = (*env)->GetFloatArrayElements(env,ju0,&isCopy);
float * v0 = (*env)->GetFloatArrayElements(env,jv0,&isCopy);
float * w0 = (*env)->GetFloatArrayElements(env,jw0,&isCopy);
vel_step(DIM_X, u, v, w, u0, v0, w0, VISCOSITY_RATE, timestep);
dens_step(DIM_X, x, x0, u, v, w, DIFFUSION_RATE, timestep);
(*env)->SetFloatArrayRegion(env,jx,0,DIM_X*DIM_X*DIM_X,x);
(*env)->SetFloatArrayRegion(env,jx0,0,DIM_X*DIM_X*DIM_X,x0);
(*env)->SetFloatArrayRegion(env,ju,0,DIM_X*DIM_X*DIM_X,u);
(*env)->SetFloatArrayRegion(env,jv,0,DIM_X*DIM_X*DIM_X,v);
(*env)->SetFloatArrayRegion(env,jw,0,DIM_X*DIM_X*DIM_X,w);
(*env)->SetFloatArrayRegion(env,ju0,0,DIM_X*DIM_X*DIM_X,u0);
(*env)->SetFloatArrayRegion(env,jv0,0,DIM_X*DIM_X*DIM_X,v0);
(*env)->SetFloatArrayRegion(env,jw0,0,DIM_X*DIM_X*DIM_X,w0);
// for(int i=0; i<DIM_X; i++){for(int j=0; j<DIM_X; j++){for(int k=0; k<DIM_X; k++){
// (*env)->SetFloatArrayRegion(env,jx,0,DIM_X*DIM_X*DIM_X,x);
// // jx[IX(i,j,k)] = x[IX(i,j,k)];
// }}}
}
int main() {
printf("Hello World from Project Panama Baeldung Article");
return 0;
}
void diffuse(int N, int b, float * x, float * x0, float diff, float dt){
float a=dt*diff*N*N*N;
lin_solve(N, b, x, x0, a, 1+6*a);
}
void advect(int N, int b, float * d, float * d0, float * u, float * v, float * w, float dt){
int i, j, k, i0, j0, k0, i1, j1, k1;
float x, y, z, s0, t0, s1, t1, u1, u0, dtx,dty,dtz;
dtx=dty=dtz=dt*N;
for ( i=1 ; i<N-1 ; i++ ) { for ( j=1 ; j<N-1 ; j++ ) { for ( k=1 ; k<N-1 ; k++ ) {
x = i-dtx*u[IX(i,j,k)]; y = j-dty*v[IX(i,j,k)]; z = k-dtz*w[IX(i,j,k)];
if (x<0.5f) x=0.5f; if (x>N+0.5f) x=N+0.5f; i0=(int)x; i1=i0+1;
if (y<0.5f) y=0.5f; if (y>N+0.5f) y=N+0.5f; j0=(int)y; j1=j0+1;
if (z<0.5f) z=0.5f; if (z>N+0.5f) z=N+0.5f; k0=(int)z; k1=k0+1;
s1 = x-i0; s0 = 1-s1; t1 = y-j0; t0 = 1-t1; u1 = z-k0; u0 = 1-u1;
if(i0 >= N){
i0 = N - 1;
}
if(j0 >= N){
j0 = N - 1;
}
if(k0 >= N){
k0 = N - 1;
}
if(i1 >= N){
i1 = N - 1;
}
if(j1 >= N){
j1 = N - 1;
}
if(k1 >= N){
k1 = N - 1;
}
d[IX(i,j,k)] = s0*(t0*u0*d0[IX(i0,j0,k0)]+t1*u0*d0[IX(i0,j1,k0)]+t0*u1*d0[IX(i0,j0,k1)]+t1*u1*d0[IX(i0,j1,k1)])+
s1*(t0*u0*d0[IX(i1,j0,k0)]+t1*u0*d0[IX(i1,j1,k0)]+t0*u1*d0[IX(i1,j0,k1)]+t1*u1*d0[IX(i1,j1,k1)]);
}}}
set_bnd(N, b, d);
}
void dens_step(int N, float * x, float * x0, float * u, float * v, float * w, float diff, float dt){
// add_source ( N, x, x0, dt );
SWAP ( x0, x ); diffuse ( N, 0, x, x0, diff, dt );
SWAP ( x0, x ); advect ( N, 0, x, x0, u, v, w, dt );
}
void vel_step(int N, float * u, float * v, float * w, float * u0, float * v0, float * w0, float visc, float dt){
// add_source ( N, u, u0, dt ); add_source ( N, v, v0, dt );
// SWAP ( u0, u ); diffuse ( N, 1, u, u0, visc, dt );
// SWAP ( v0, v ); diffuse ( N, 2, v, v0, visc, dt );
// project ( N, u, v, w, u0, v0 );
// SWAP ( u0, u ); SWAP ( v0, v );
// advect ( N, 1, u, u0, u0, v0, dt ); advect ( N, 2, v, v0, u0, v0, dt );
// project ( N, u, v, w, u0, v0 );
// add_source(N, v, x, dt);
SWAP(u0, u);
diffuse(N, 1, u, u0, visc, dt);
SWAP(v0, v);
diffuse(N, 2, v, v0, visc, dt);
SWAP(w0, w);
diffuse(N, 3, w, w0, visc, dt);
project(N, u, v, w, u0, v0);
SWAP(u0, u);
SWAP(v0, v);
SWAP(w0, w);
advect(N, 1, u, u0, u0, v0, w0, dt);
advect(N, 2, v, v0, u0, v0, w0, dt);
advect(N, 3, w, w0, u0, v0, w0, dt);
project(N, u, v, w, u0, v0);
}
void project(int N, float * u, float * v, float * w, float * p, float * div){
int i, j, k;
for ( i=1 ; i<N-1 ; i++ ) { for ( j=1 ; j<N-1 ; j++ ) { for ( k=1 ; k<N-1 ; k++ ) {
div[IX(i,j,k)] = (float)(-1.0/3.0*((u[IX(i+1,j,k)]-u[IX(i-1,j,k)])/N+(v[IX(i,j+1,k)]-v[IX(i,j-1,k)])/N+(w[IX(i,j,k+1)]-w[IX(i,j,k-1)])/N));
p[IX(i,j,k)] = 0;
}}}
set_bnd(N, 0, div);
set_bnd(N, 0, p);
lin_solve(N, 0, p, div, 1, 6);
for ( i=1 ; i<N-1 ; i++ ) { for ( j=1 ; j<N-1 ; j++ ) { for ( k=1 ; k<N-1 ; k++ ) {
u[IX(i,j,k)] -= 0.5f*N*(p[IX(i+1,j,k)]-p[IX(i-1,j,k)]);
v[IX(i,j,k)] -= 0.5f*N*(p[IX(i,j+1,k)]-p[IX(i,j-1,k)]);
w[IX(i,j,k)] -= 0.5f*N*(p[IX(i,j,k+1)]-p[IX(i,j,k-1)]);
}}}
set_bnd(N, 1, u);
set_bnd(N, 2, v);
set_bnd(N, 3, w);
}
void lin_solve(int N, int b, float* x, float* x0, float a, float c){
int i, j, k, l;
// iterate the solver
for ( l=0 ; l<LINEARSOLVERTIMES ; l++ ) {
// update for each cell
for ( i=1 ; i<N-1 ; i++ ) { for ( j=1 ; j<N-1 ; j++ ) { for ( k=1 ; k<N-1 ; k++ ) {
x[IX(i,j,k)] = (x0[IX(i,j,k)] + a*(x[IX(i-1,j,k)]+x[IX(i+1,j,k)]+x[IX(i,j-1,k)]+x[IX(i,j+1,k)]+x[IX(i,j,k-1)]+x[IX(i,j,k+1)]))/c;
}}}
set_bnd(N, b, x);
}
}
// void set_bnd ( int N, int b, float * x ){
// int i;
// // for ( i=1 ; i<=N ; i++ ) {
// // x[IX(0 ,i)] = b==1 ? -x[IX(1,i)] : x[IX(1,i)];
// // x[IX(N+1,i)] = b==1 ? -x[IX(N,i)] : x[IX(N,i)];
// // x[IX(i,0 )] = b==2 ? -x[IX(i,1)] : x[IX(i,1)];
// // x[IX(i,N+1)] = b==2 ? -x[IX(i,N)] : x[IX(i,N)];
// // }
// // x[IX(0 ,0 )] = 0.5*(x[IX(1,0 )]+x[IX(0 ,1)]);
// // x[IX(0 ,N+1)] = 0.5*(x[IX(1,N+1)]+x[IX(0 ,N )]);
// // x[IX(N+1,0 )] = 0.5*(x[IX(N,0 )]+x[IX(N+1,1)]);
// // x[IX(N+1,N+1)] = 0.5*(x[IX(N,N+1)]+x[IX(N+1,N )]);
// }
void set_bnd(int N, int b, float * target){
int DIM = N;
for(int x=1; x < DIM-1; x++){
for(int y = 1; y < DIM-1; y++){
//((x)+(DIM)*(y) + (DIM)*(DIM)*(z))
target[0 + DIM * x + DIM * DIM * y] = b==1 ? -target[1 + DIM * x + DIM * DIM * y] : target[1 + DIM * x + DIM * DIM * y];
target[IX(DIM-1,x,y)] = b==1 ? -target[IX(DIM-2,x,y)] : target[IX(DIM-2,x,y)];
target[IX(x,0,y)] = b==2 ? -target[IX(x,1,y)] : target[IX(x,1,y)];
target[IX(x,DIM-1,y)] = b==2 ? -target[IX(x,DIM-2,y)] : target[IX(x,DIM-2,y)];
target[IX(x,y,0)] = b==3 ? -target[IX(x,y,1)] : target[IX(x,y,1)];
target[IX(x,y,DIM-1)] = b==3 ? -target[IX(x,y,DIM-2)] : target[IX(x,y,DIM-2)];
}
}
for(int x = 1; x < DIM-1; x++){
target[IX(x,0,0)] = (float)(0.5f * (target[IX(x,1,0)] + target[IX(x,0,1)]));
target[IX(x,DIM-1,0)] = (float)(0.5f * (target[IX(x,DIM-2,0)] + target[IX(x,DIM-1,1)]));
target[IX(x,0,DIM-1)] = (float)(0.5f * (target[IX(x,1,DIM-1)] + target[IX(x,0,DIM-2)]));
target[IX(x,DIM-1,DIM-1)] = (float)(0.5f * (target[IX(x,DIM-2,DIM-1)] + target[IX(x,DIM-1,DIM-2)]));
target[IX(0,x,0)] = (float)(0.5f * (target[IX(1,x,0)] + target[IX(0,x,1)]));
target[IX(DIM-1,x,0)] = (float)(0.5f * (target[IX(DIM-2,x,0)] + target[IX(DIM-1,x,1)]));
target[IX(0,x,DIM-1)] = (float)(0.5f * (target[IX(1,x,DIM-1)] + target[IX(0,x,DIM-2)]));
target[IX(DIM-1,x,DIM-1)] = (float)(0.5f * (target[IX(DIM-2,x,DIM-1)] + target[IX(DIM-1,x,DIM-2)]));
target[IX(0,0,x)] = (float)(0.5f * (target[IX(1,0,x)] + target[IX(0,1,x)]));
target[IX(DIM-1,0,x)] = (float)(0.5f * (target[IX(DIM-2,0,x)] + target[IX(DIM-1,1,x)]));
target[IX(0,DIM-1,x)] = (float)(0.5f * (target[IX(1,DIM-1,x)] + target[IX(0,DIM-2,x)]));
target[IX(DIM-1,DIM-1,x)] = (float)(0.5f * (target[IX(DIM-2,DIM-1,x)] + target[IX(DIM-1,DIM-2,x)]));
}
target[IX(0,0,0)] = (float)((target[IX(1,0,0)]+target[IX(0,1,0)]+target[IX(0,0,1)])/3.0);
target[IX(DIM-1,0,0)] = (float)((target[IX(DIM-2,0,0)]+target[IX(DIM-1,1,0)]+target[IX(DIM-1,0,1)])/3.0);
target[IX(0,DIM-1,0)] = (float)((target[IX(1,DIM-1,0)]+target[IX(0,DIM-2,0)]+target[IX(0,DIM-1,1)])/3.0);
target[IX(0,0,DIM-1)] = (float)((target[IX(0,0,DIM-2)]+target[IX(1,0,DIM-1)]+target[IX(0,1,DIM-1)])/3.0);
target[IX(DIM-1,DIM-1,0)] = (float)((target[IX(DIM-2,DIM-1,0)]+target[IX(DIM-1,DIM-2,0)]+target[IX(DIM-1,DIM-1,1)])/3.0);
target[IX(0,DIM-1,DIM-1)] = (float)((target[IX(1,DIM-1,DIM-1)]+target[IX(0,DIM-2,DIM-1)]+target[IX(0,DIM-1,DIM-2)])/3.0);
target[IX(DIM-1,0,DIM-1)] = (float)((target[IX(DIM-1,0,DIM-2)]+target[IX(DIM-2,0,DIM-1)]+target[IX(DIM-1,1,DIM-1)])/3.0);
target[IX(DIM-1,DIM-1,DIM-1)] = (float)((target[IX(DIM-1,DIM-1,DIM-2)]+target[IX(DIM-1,DIM-2,DIM-1)]+target[IX(DIM-1,DIM-1,DIM-2)])/3.0);
}

31
src/main/c/includes/electrosphere_FluidSim.h Normal file
View File

@ -0,0 +1,31 @@
/* DO NOT EDIT THIS FILE - it is machine generated */
#include <jni.h>
/* Header for class electrosphere_FluidSim */
#ifndef _Included_electrosphere_FluidSim
#define _Included_electrosphere_FluidSim
#ifdef __cplusplus
extern "C" {
#endif
#undef electrosphere_FluidSim_DIM
#define electrosphere_FluidSim_DIM 18L
#undef electrosphere_FluidSim_DIFFUSION_CONSTANT
#define electrosphere_FluidSim_DIFFUSION_CONSTANT 1.0E-4f
#undef electrosphere_FluidSim_VISCOSITY_CONSTANT
#define electrosphere_FluidSim_VISCOSITY_CONSTANT 1.0E-4f
#undef electrosphere_FluidSim_LINEARSOLVERTIMES
#define electrosphere_FluidSim_LINEARSOLVERTIMES 20L
#undef electrosphere_FluidSim_GRAVITY
#define electrosphere_FluidSim_GRAVITY -10000.0f
/*
* Class: electrosphere_FluidSim
* Method: simulate
* Signature: (III[F[F[F[F[F[F[F[FFFF)V
*/
JNIEXPORT void JNICALL Java_electrosphere_FluidSim_simulate
(JNIEnv *, jobject, jint, jint, jint, jfloatArray, jfloatArray, jfloatArray, jfloatArray, jfloatArray, jfloatArray, jfloatArray, jfloatArray, jfloat, jfloat, jfloat);
#ifdef __cplusplus
}
#endif
#endif

128
src/main/java/electrosphere/FluidSim.java
View File

@ -9,36 +9,21 @@ import java.util.List;
import java.util.Random; import java.util.Random;
import java.util.Set; import java.util.Set;
import jdk.incubator.foreign.MemorySegment;
import jdk.incubator.vector.FloatVector;
import jdk.incubator.vector.VectorMask;
import jdk.incubator.vector.VectorOperators;
import jdk.incubator.vector.VectorOperators.Associative;
import org.joml.Vector2i; import org.joml.Vector2i;
import org.lwjgl.BufferUtils; import org.lwjgl.BufferUtils;
import org.lwjgl.PointerBuffer; import org.lwjgl.PointerBuffer;
import org.lwjgl.opencl.CL30;
import org.lwjgl.system.MemoryUtil; import org.lwjgl.system.MemoryUtil;
import electrosphere.opencl.CLBuffer;
import electrosphere.opencl.CLCommandQueue;
import electrosphere.opencl.CLContext;
import electrosphere.opencl.CLDevice;
import electrosphere.opencl.CLEvent;
import electrosphere.opencl.CLImage;
import electrosphere.opencl.CLKernel;
import electrosphere.opencl.CLPlatform;
import electrosphere.opencl.CLProgram;
import electrosphere.opencl.CLReadImageResult;
import electrosphere.opencl.CLUtilities;
import electrosphere.opencl.LWJGLContext;
/** /**
* Simulates a fluid via opencl * Simulates a fluid via opencl
*/ */
public class FluidSim { public class FluidSim {
static {
System.out.println(System.getProperty("user.dir"));
System.load(System.getProperty("user.dir") + "/shared-folder/libfluidsim.dll");
}
public static final int DIM = 18; public static final int DIM = 18;
float[] x = new float[DIM * DIM * DIM]; float[] x = new float[DIM * DIM * DIM];
@ -118,10 +103,11 @@ public class FluidSim {
} }
} }
simulate(DIM, DIM, DIM, x, x0, u, v, w, u0, v0, w0, DIFFUSION_CONSTANT, VISCOSITY_CONSTANT, timestep);
vel_step(DIM, DIM, DIM, u, v, w, u0, v0, w0, VISCOSITY_CONSTANT, timestep); // vel_step(DIM, DIM, DIM, u, v, w, u0, v0, w0, VISCOSITY_CONSTANT, timestep);
dens_step(DIM, DIM, DIM, x, x0, u, v, w, DIFFUSION_CONSTANT, timestep); // dens_step(DIM, DIM, DIM, x, x0, u, v, w, DIFFUSION_CONSTANT, timestep);
sum(); sum();
} }
@ -144,7 +130,7 @@ public class FluidSim {
void diffuse(int M, int N, int O, int b, float[] x, float[] x0, float diff, float dt){ void diffuse(int M, int N, int O, int b, float[] x, float[] x0, float diff, float dt){
int max = Math.max(Math.max(M, N), Math.max(N, O)); int max = Math.max(Math.max(M, N), Math.max(N, O));
float a=dt*diff*max*max*max; float a=dt*diff*max*max*max;
linSolveJacobian(M, N, O, b, x, x0, a, 1+6*a); lin_solve(M, N, O, b, x, x0, a, 1+6*a);
} }
void setBounds(float[] target, int b){ void setBounds(float[] target, int b){
@ -245,7 +231,7 @@ public class FluidSim {
setBounds(div, 0); setBounds(div, 0);
setBounds(p, 0); setBounds(p, 0);
linSolveJacobian(M, N, O, 0, p, div, 1, 6); lin_solve(M, N, O, 0, p, div, 1, 6);
for ( i=1 ; i<M-1 ; i++ ) { for ( j=1 ; j<N-1 ; j++ ) { for ( k=1 ; k<O-1 ; k++ ) { for ( i=1 ; i<M-1 ; i++ ) { for ( j=1 ; j<N-1 ; j++ ) { for ( k=1 ; k<O-1 ; k++ ) {
u[IX(i,j,k)] -= 0.5f*M*(p[IX(i+1,j,k)]-p[IX(i-1,j,k)]); u[IX(i,j,k)] -= 0.5f*M*(p[IX(i+1,j,k)]-p[IX(i-1,j,k)]);
@ -309,6 +295,23 @@ public class FluidSim {
// System.out.println(sum[0] + " " + sum[1] + " " + sum[2] + " " + sum[3]); // System.out.println(sum[0] + " " + sum[1] + " " + sum[2] + " " + sum[3]);
} }
private native void simulate(
int DIM_X,
int DIM_Y,
int DIM_Z,
float[] x,
float[] x0,
float[] u,
float[] v,
float[] w,
float[] u0,
float[] v0,
float[] w0,
float DIFFUSION_CONSTANT,
float VISCOSITY_CONSTANT,
float timestep
);
public float[] getData(){ public float[] getData(){
return x; return x;
} }
@ -344,81 +347,4 @@ public class FluidSim {
} }
} }
static final int LANE_WIDTH = 8;
void linSolveJacobian(int M, int N, int O, int b, float[] x, float x0[], float a, float c){
//x contains the solution to the problem
//x0 contains A
// float[] a = new float[]{
// 1,2,3,4,5,6,7,8,9,10,11,12,13,14,15,16
// };
// FloatVector vec = FloatVector.fromArray(FloatVector.SPECIES_128, a, 0);
// FloatVector finalVec = vec.add(3);
// float[] b = finalVec.toArray();
// for(int i = 0; i < b.length; i++){
// System.out.println(b[i]);
// }
float[] vectorArrView;
float[] srcArr, destArr;
for(int k = 0; k < LINEARSOLVERTIMES; k++){
if(k / 2 == 0){
srcArr = x;
destArr = jacobiAltArray;
} else {
srcArr = jacobiAltArray;
destArr = x;
}
for(int i = 0; i < vectorOffsetMap.length; i += LANE_WIDTH){
if(i + LANE_WIDTH < vectorOffsetMap.length){
// for ( i=1 ; i<M-1 ; i++ ) { for ( j=1 ; j<N-1 ; j++ ) { for ( k=1 ; k<O-1 ; k++ ) {
// x[IX(i,j,k)] = (x0[IX(i,j,k)] + a*(x[IX(i-1,j,k)]+x[IX(i+1,j,k)]+x[IX(i,j-1,k)]+x[IX(i,j+1,k)]+x[IX(i,j,k-1)]+x[IX(i,j,k+1)]))/c;
// }}}
// adderVector = FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapXN,i);
// x0Vector = FloatVector.fromArray(FloatVector.SPECIES_256,x0,0,vectorOffsetMap,i);
vectorArrView = FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapXN,i)
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapXP,i))
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapYN,i))
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapYP,i))
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapZN,i))
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapZP,i))
.mul(a)
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x0,0,vectorOffsetMap,i))
.div(c).toArray();
for(int l = 0; l < LANE_WIDTH; l++){
x[vectorOffsetMap[i+l]] = vectorArrView[l];
}
} else {
int numLeft = vectorOffsetMap.length - i;
VectorMask<Float> mask = FloatVector.SPECIES_256.indexInRange(O, numLeft);
// for ( i=1 ; i<M-1 ; i++ ) { for ( j=1 ; j<N-1 ; j++ ) { for ( k=1 ; k<O-1 ; k++ ) {
// x[IX(i,j,k)] = (x0[IX(i,j,k)] + a*(x[IX(i-1,j,k)]+x[IX(i+1,j,k)]+x[IX(i,j-1,k)]+x[IX(i,j+1,k)]+x[IX(i,j,k-1)]+x[IX(i,j,k+1)]))/c;
// }}}
// adderVector = FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapXN,i,mask);
// x0Vector = FloatVector.fromArray(FloatVector.SPECIES_256,x0,0,vectorOffsetMap,i,mask);
vectorArrView = FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapXN,i,mask)
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapXP,i))
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapYN,i))
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapYP,i))
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapZN,i))
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x,0,vectorOffsetMapZP,i))
.mul(a)
.add(FloatVector.fromArray(FloatVector.SPECIES_256,x0,0,vectorOffsetMap,i,mask))
.div(c).toArray();
for(int l = 0; l < numLeft; l++){
x[vectorOffsetMap[i+l]] = vectorArrView[l];
}
}
}
setBounds(destArr, b);
}
//x[IX(i,j,k)] = (x0[IX(i,j,k)] + a*(x[IX(i-1,j,k)]+x[IX(i+1,j,k)]+x[IX(i,j-1,k)]+x[IX(i,j+1,k)]+x[IX(i,j,k-1)]+x[IX(i,j,k+1)]))/c;
}
} }

46
src/main/java/electrosphere/Main.java
View File

@ -15,28 +15,32 @@ import electrosphere.render.Mesh;
public class Main { public class Main {
public static void main(String args[]){ public static void main(String args[]){
GLFWContext.init(); try {
FluidSim sim = new FluidSim(); GLFWContext.init();
sim.setup(); FluidSim sim = new FluidSim();
Mesh.meshInitially(sim); sim.setup();
int i = 0; Mesh.meshInitially(sim);
long time = 0; int i = 0;
long lastTime = 0; long time = 0;
while(true){ long lastTime = 0;
try { while(true){
TimeUnit.MILLISECONDS.sleep(1); try {
} catch (InterruptedException e) { TimeUnit.MILLISECONDS.sleep(2);
e.printStackTrace(); } catch (InterruptedException e) {
} e.printStackTrace();
lastTime = System.currentTimeMillis(); }
sim.simulate(i,0.001f); lastTime = System.currentTimeMillis();
time = time + (System.currentTimeMillis() - lastTime); sim.simulate(i,0.001f);
Mesh.remesh(sim); time = time + (System.currentTimeMillis() - lastTime);
GLFWContext.redraw(); Mesh.remesh(sim);
i++; GLFWContext.redraw();
if(i == 1000){ i++;
System.out.println(time / 1000.0); if(i == 1000){
System.out.println(time / 1000.0);
}
} }
} catch(Throwable ex){
ex.printStackTrace(System.err);
} }
} }

89
src/main/java/electrosphere/MyTest.java
View File

@ -1,89 +0,0 @@
package electrosphere;
import java.io.File;
import java.io.IOException;
import java.nio.ByteBuffer;
import java.nio.IntBuffer;
import java.nio.file.Files;
import java.util.ArrayList;
import java.util.Collections;
import java.util.LinkedList;
import java.util.List;
import java.util.Set;
import org.joml.Vector2i;
import org.lwjgl.*;
import org.lwjgl.glfw.*;
import org.lwjgl.opencl.*;
import org.lwjgl.system.*;
import electrosphere.opencl.CLBuffer;
import electrosphere.opencl.CLCommandQueue;
import electrosphere.opencl.CLContext;
import electrosphere.opencl.CLDevice;
import electrosphere.opencl.CLEvent;
import electrosphere.opencl.CLImage;
import electrosphere.opencl.CLKernel;
import electrosphere.opencl.CLPlatform;
import electrosphere.opencl.CLProgram;
import electrosphere.opencl.CLReadImageResult;
import electrosphere.opencl.CLUtilities;
import electrosphere.opencl.LWJGLContext;
import static org.lwjgl.glfw.Callbacks.*;
import static org.lwjgl.glfw.GLFW.*;
import static org.lwjgl.opencl.CL10.*;
import static org.lwjgl.opencl.KHRICD.*;
import static org.lwjgl.system.MemoryStack.*;
import static org.lwjgl.system.MemoryUtil.*;
import static org.lwjgl.glfw.GLFWNativeGLX.*;
import static org.lwjgl.glfw.GLFWNativeWGL.*;
import static org.lwjgl.glfw.GLFWNativeX11.*;
import static org.lwjgl.opencl.CL10GL.*;
import static org.lwjgl.opencl.KHRGLSharing.*;
import static org.lwjgl.opengl.ARBCLEvent.*;
import static org.lwjgl.opengl.CGL.*;
import static org.lwjgl.opengl.GL30C.*;
import static org.lwjgl.opengl.WGL.*;
public class MyTest {
//Good reference: https://www.nersc.gov/assets/pubs_presos/MattsonTutorialSC14.pdf
//Khronos api reference cheatsheet: https://www.khronos.org/files/opencl30-reference-guide.pdf
//error code reference: https://streamhpc.com/blog/2013-04-28/opencl-error-codes/
//relevant: https://blog.lwjgl.org/memory-management-in-lwjgl-3/
//be mindful of: https://www.baeldung.com/java-memory-leaks
public static void run(){
FluidSim sim = new FluidSim();
sim.setup();
double deltaTime = 0.0f;
double lastFrame = 0.0f;
double frameRate = 0.0f;
GLFW.glfwInit();
int iterations = 0;
while(true){
sim.simulate(1, 0.2f);
//framerate calculations
double currentTime = glfwGetTime();
deltaTime = currentTime - lastFrame;
frameRate = 1.0 / deltaTime;
lastFrame = currentTime;
iterations++;
// System.out.print("\r" + currentTime + " " + frameRate + " _ " + iterations + " _ " + Runtime.getRuntime().totalMemory() / 1024 / 1024);
}
}
}

35
src/main/java/electrosphere/opencl/CLBuffer.java
View File

@ -1,35 +0,0 @@
package electrosphere.opencl;
import java.nio.IntBuffer;
import org.lwjgl.opencl.CL30;
/**
* Wrapper object for an opencl buffer
*/
public class CLBuffer {
//The pointer to the buffer
long pointer;
/**
* Creates a opencl buffer with a specified size
* @param context The context to create the buffer under
* @param flags The flags for the buffer
* @param size The size of the buffer
*/
protected CLBuffer(CLContext context, long flags, long size){
pointer = CL30.clCreateBuffer(context.getPointer(), flags, size, context.getErrorCodeBuffer());
}
/**
* Creates an opencl buffer based on a provided buffer
* @param context The context to create the buffer under
* @param flags The flags for the buffer
* @param hostBuffer The cpu buffer to base the opencl buffer off of
*/
protected CLBuffer(CLContext context, long flags, IntBuffer hostBuffer){
pointer = CL30.clCreateBuffer(context.getPointer(), flags, hostBuffer, context.getErrorCodeBuffer());
}
}

434
src/main/java/electrosphere/opencl/CLCommandQueue.java
View File

@ -1,434 +0,0 @@
package electrosphere.opencl;
import java.nio.ByteBuffer;
import java.util.List;
import org.joml.Vector2i;
import org.lwjgl.BufferUtils;
import org.lwjgl.PointerBuffer;
import org.lwjgl.opencl.CL30;
import org.lwjgl.system.MemoryStack;
import org.lwjgl.system.MemoryUtil;
/**
* Wrapper object for an opencl command queue
*/
public class CLCommandQueue {
//The lwjgl context
LWJGLContext lwjglContext;
//The pointer to the command queue
long pointer;
/**
* Creates a command queue for a provided context
* @param context The context
*/
protected CLCommandQueue(LWJGLContext lwjglContext, CLContext clContext){
this.lwjglContext = lwjglContext;
pointer = CL30.clCreateCommandQueue(clContext.getPointer(), clContext.getDevice().getId(), MemoryUtil.NULL, clContext.getErrorCodeBuffer());
}
/**
* Gets the pointer for this command queue
* @return The pointer as a long
*/
public long getPointer(){
return pointer;
}
/**
* Queues a 2d kernel to work
* @param kernel The kernel
* @param waitList The list of events to wait on
* @param workersInEachDimension The workers in each dimension of the data
* @param workersInEachGroup The workers in each group of workers
* @return The event representing this queued work
*/
public CLEvent queueKernelWork(
CLKernel kernel,
List<CLEvent> waitList,
Vector2i workersInEachDimension,
Vector2i workersInEachGroup
){
long[] dimensionArray = new long[]{
workersInEachDimension.x,
workersInEachDimension.y
};
long[] groupArray = new long[]{
workersInEachGroup.x,
workersInEachGroup.y
};
//dimensionality must be 2 as the work arrays are 2D vectors
CLEvent event = queueKernelWork(kernel,waitList,2,dimensionArray,groupArray);
return event;
}
/**
* Queues kernel work with a default worker offset vector
* @param kernel The kernel to queue work on
* @param waitList The waitlist of events to wait for before executing this work
* @param workDimensions The dimensionality of the data to act upon
* @param workersInEachDimension The total number of workers in each dimension
* @param workersInEachGroup The total number of workers in each work group
* @return The opencl event for this work item
*/
public CLEvent queueKernelWork(
CLKernel kernel,
List<CLEvent> waitList,
int workDimensions,
long[] workersInEachDimension,
long[] workersInEachGroup
){
return queueKernelWork(kernel, waitList, workDimensions, null, workersInEachDimension, workersInEachGroup);
}
/**
* Queues work on a given opencl kernel
* @param kernel The kernel to queue work on
* @param waitList The waitlist of events to wait for before executing this work
* @param workDimensions The dimensionality of the data to act upon
* @param globalWorkerOffsetVector The vector specifying the amount of offset per dimension between given global workers
* @param workersInEachDimension The total number of workers in each dimension
* @param workersInEachGroup The total number of workers in each work group
* @return The opencl event for this work item
*/
public CLEvent queueKernelWork(
CLKernel kernel,
List<CLEvent> waitList,
int workDimensions,
long[] globalWorkerOffsetVector, //basically a vector that describes the increment in each dimension. Unlikely to see use generally
long[] workersInEachDimension, //basically the number of sub arrays
long[] workersInEachGroup // basically the size of each subarray to work one
){
CLEvent rVal = null;
try(MemoryStack stack = MemoryStack.stackPush()){
//
//construct and fill buffers for opencl call
//global worker offsets
PointerBuffer globalWorkerOffset = null;
if(globalWorkerOffsetVector != null){
globalWorkerOffset = stack.mallocPointer(workDimensions);
globalWorkerOffset.put(globalWorkerOffsetVector);
globalWorkerOffset.flip();
}
//global worker size
PointerBuffer globalWorkerSize = null;
globalWorkerSize = stack.mallocPointer(workDimensions);
globalWorkerSize.put(workersInEachDimension);
globalWorkerSize.flip();
//local worker size
PointerBuffer localWorkerSize = stack.mallocPointer(workDimensions);
localWorkerSize.put(workersInEachGroup);
localWorkerSize.flip();
//the events to wait for before executing this one
PointerBuffer eventWaitList = null;
if(waitList != null && waitList.size() > 0){
eventWaitList = stack.mallocPointer(waitList.size());
for(CLEvent toWaitEvent : waitList){
eventWaitList.put(toWaitEvent.getPointer());
}
}
PointerBuffer event = stack.mallocPointer(1);
//
//queue work
CL30.clEnqueueNDRangeKernel(
this.pointer,
kernel.getPointer(),
workDimensions,
globalWorkerOffset,
globalWorkerSize,
localWorkerSize,
eventWaitList,
event
);
//create returned event
rVal = new CLEvent(event.get());
}
return rVal;
}
/**
* Sets an argument for a given kernel
* @param kernel The kernel
* @param argumentIndex The index of the argument (0-index'd)
* @param argumentValue The value to set the argument to
* @return The result of the argument set operation
*/
public int setKernelArgument(CLKernel kernel, int argumentIndex, long argumentValue){
return CL30.clSetKernelArg(kernel.getPointer(),argumentIndex,argumentValue);
}
/**
* Sets an argument for a given kernel to be equal to a given image
* @param kernel The kernel
* @param argumentIndex The index of the argument (0-index'd)
* @param image The value to set the argument to
* @return The result of the argument set operation
*/
public int setKernelArgument(CLKernel kernel, int argumentIndex, CLImage image){
int rVal = 0;
try(MemoryStack stack = MemoryStack.stackPush()){
PointerBuffer pointerBuffer = stack.mallocPointer(1);
pointerBuffer.put(image.getPointer());
pointerBuffer.flip();
rVal = CL30.clSetKernelArg(kernel.getPointer(),argumentIndex,pointerBuffer);
}
return rVal;
}
/**
* Reads an image from opencl global memory
* @param image The image
* @param waitList The wait list of events to make this query pend on
* @return The byte buffer containing the data of the image
*/
public CLReadImageResult readImage(CLImage image, List<CLEvent> waitList){
CLReadImageResult rVal = null;
//the cpu buffer to write into from gpu memory
int bufferSize = (int)(image.getWidth() * image.getByteWidth());
if(image.getHeight() > 0){
bufferSize = (int)(bufferSize * image.getHeight());
}
if(image.getDepth() > 0){
bufferSize = (int)(bufferSize * image.getDepth());
}
ByteBuffer outputBuffer = MemoryUtil.memAlloc(bufferSize);
try(MemoryStack stack = MemoryStack.stackPush()){
boolean blocking = false; //should the call block execution until the image is read?
//defines the offset in outputBuffer to start writing (should be a vec)
PointerBuffer origin = stack.mallocPointer(3);
origin.put(0);
origin.put(0);
origin.put(0);
origin.flip();
//The dimensions of the image
PointerBuffer region = stack.mallocPointer(3);
region.put(image.getWidth());
if(image.getHeight() == 0){
region.put(1); //If it's a 1D image, this should be set to 1
} else {
region.put(image.getHeight());
}
if(image.getDepth() == 0){
region.put(1); //If it's a 2D image, this should be set to 1
} else {
region.put(image.getDepth());
}
region.flip();
//how many bytes make up 1 line of the data
long rowPitch = image.getByteWidth() * image.getWidth();
//how many bytes make up a single 2d plane of pixels. Should be 0 if image is 1d or 2d
long slicePitch = 0;
if(image.getDepth() > 0){
slicePitch = image.getByteWidth() * image.getWidth() * image.getHeight();
}
//the event management buffers
PointerBuffer eventWaitList = null;
if(waitList != null && waitList.size() > 0){
eventWaitList = stack.mallocPointer(waitList.size());
for(CLEvent toWaitEvent : waitList){
eventWaitList.put(toWaitEvent.getPointer());
}
eventWaitList.flip();
}
PointerBuffer event = stack.mallocPointer(1);
//actually send to opencl
int errorCode = CL30.clEnqueueReadImage(
this.getPointer(),
image.getPointer(),
blocking,
origin,
region,
rowPitch,
slicePitch,
outputBuffer,
eventWaitList,
event
);
//error check
CLUtilities.checkCLError(errorCode);
//construct return object
CLEvent returnEvent = new CLEvent(event.get());
rVal = new CLReadImageResult(outputBuffer,returnEvent);
}
return rVal;
}
/**
* Writes an array of bytes to a device-side image in opencl
* @param image The image to write to
* @param data The data to write
* @param waitList The waitlist of events to pend this write operation on
* @return The CLEvent encapsulating this write operation
*/
public CLEvent writeToDeviceImage(CLImage image, List<CLEvent> waitList, int[] data){
CLEvent rVal = null;
try(MemoryStack stack = MemoryStack.stackPush()){
boolean blocking = true;
//defines the offset in outputBuffer to start writing (should be a vec)
PointerBuffer origin = stack.mallocPointer(3);
origin.put(0);
origin.put(0);
origin.put(0);
origin.flip();
//The dimensions of the image
PointerBuffer region = stack.mallocPointer(3);
region.put(image.getWidth());
if(image.getHeight() == 0){
region.put(1); //If it's a 1D image, this should be set to 1
} else {
region.put(image.getHeight());
}
if(image.getDepth() == 0){
region.put(1); //If it's a 2D image, this should be set to 1
} else {
region.put(image.getDepth());
}
region.flip();
//how many bytes make up 1 line of the data
long rowPitch = image.getByteWidth() * image.getWidth();
//how many bytes make up a single 2d plane of pixels. Should be 0 if image is 1d or 2d
long slicePitch = 0;
if(image.getDepth() > 0){
slicePitch = image.getByteWidth() * image.getWidth() * image.getHeight();
}
//the event management buffers
PointerBuffer eventWaitList = null;
if(waitList != null && waitList.size() > 0){
eventWaitList = stack.mallocPointer(waitList.size());
for(CLEvent toWaitEvent : waitList){
eventWaitList.put(toWaitEvent.getPointer());
}
eventWaitList.flip();
}
PointerBuffer event = stack.mallocPointer(1);
int errorCode = CL30.clEnqueueWriteImage(this.getPointer(), image.getPointer(), blocking, origin, region, rowPitch, slicePitch, data, eventWaitList, event);
//error check
CLUtilities.checkCLError(errorCode);
//construct return event
rVal = new CLEvent(event.get());
}
return rVal;
}
/**
* Writes a buffer of bytes to a device-side image in opencl
* @param image The image to write to
* @param data The data to write
* @param waitList The waitlist of events to pend this write operation on
* @return The CLEvent encapsulating this write operation
*/
public CLEvent writeToDeviceImage(CLImage image, List<CLEvent> waitList, ByteBuffer data){
CLEvent rVal = null;
try(MemoryStack stack = MemoryStack.stackPush()){
boolean blocking = false;
//defines the offset in outputBuffer to start writing (should be a vec)
PointerBuffer origin = stack.mallocPointer(3);
origin.put(0);
origin.put(0);
origin.put(0);
origin.flip();
//The dimensions of the image
PointerBuffer region = stack.mallocPointer(3);
region.put(image.getWidth());
if(image.getHeight() == 0){
region.put(1); //If it's a 1D image, this should be set to 1
} else {
region.put(image.getHeight());
}
if(image.getDepth() == 0){
region.put(1); //If it's a 2D image, this should be set to 1
} else {
region.put(image.getDepth());
}
region.flip();
//how many bytes make up 1 line of the data
long rowPitch = image.getByteWidth() * image.getWidth();
//how many bytes make up a single 2d plane of pixels. Should be 0 if image is 1d or 2d
long slicePitch = 0;
if(image.getDepth() > 0){
slicePitch = image.getByteWidth() * image.getWidth() * image.getHeight();
}
//the event management buffers
PointerBuffer eventWaitList = null;
if(waitList != null && waitList.size() > 0){
eventWaitList = stack.mallocPointer(waitList.size());
for(CLEvent toWaitEvent : waitList){
eventWaitList.put(toWaitEvent.getPointer());
}
eventWaitList.flip();
}
PointerBuffer event = stack.mallocPointer(1);
int errorCode = CL30.clEnqueueWriteImage(this.getPointer(), image.getPointer(), blocking, origin, region, rowPitch, slicePitch, data, eventWaitList, event);
//error check
CLUtilities.checkCLError(errorCode);
//construct return event
rVal = new CLEvent(event.get());
}
return rVal;
}
/**
* Blocks thread execution until the event list has completed or otherwise resolved
* @param eventList The event list
*/
public void waitForEventList(List<CLEvent> waitList){
try(MemoryStack stack = MemoryStack.stackPush()){
PointerBuffer eventWaitList = null;
if(waitList != null && waitList.size() > 0){
eventWaitList = stack.mallocPointer(waitList.size());
for(CLEvent toWaitEvent : waitList){
eventWaitList.put(toWaitEvent.getPointer());
}
eventWaitList.flip();
}
//wait
int errorCode = CL30.clWaitForEvents(eventWaitList);
//error check
CLUtilities.checkCLError(errorCode);
}
}
}

164
src/main/java/electrosphere/opencl/CLContext.java
View File

@ -1,164 +0,0 @@
package electrosphere.opencl;
import java.nio.ByteBuffer;
import java.nio.IntBuffer;
import org.lwjgl.BufferUtils;
import org.lwjgl.PointerBuffer;
import org.lwjgl.opencl.CL30;
import org.lwjgl.opencl.CLContextCallback;
import org.lwjgl.system.MemoryUtil;
/**
* Wrapper object for an opencl context
*/
public class CLContext {
//the context pointer
long pointer;
//the context callback
CLContextCallback clContextCB;
//the error code buffer
IntBuffer errorCodeBuffer = BufferUtils.createIntBuffer(1);
int[] errorCodeArray = new int[1];
//the device this context is created on
CLDevice device;
//creates an opencl context
public CLContext(CLDevice device, PointerBuffer ctxProps){
pointer = CL30.clCreateContext(ctxProps, device.getId(), clContextCB = CLContextCallback.create(
(errinfo, private_info, cb, user_data) -> System.out.println(String.format("clCreateContext\n\tInfo: %s", MemoryUtil.memUTF8(errinfo)))
), MemoryUtil.NULL, errorCodeBuffer);
this.device = device;
}
/**
* Gets the pointer for this context
* @return The pointer
*/
public long getPointer(){
return pointer;
}
/**
* Gets the opencl error code storage buffer
* @return The opencl error code storage buffer
*/
protected IntBuffer getErrorCodeBuffer(){
return errorCodeBuffer;
}
/**
* Gets an array used for storing error codes in relevant api calls
* @return The array
*/
protected int[] getErrorCodeArray(){
return errorCodeArray;
}
/**
* Creates an opencl buffer with a specified size
* @param flags The flags for the buffer
* @param size The size of the opencl buffer in bytes
* @return The opencl buffer
*/
public CLBuffer createBuffer(long flags, long size){
CLBuffer buffer = new CLBuffer(this, flags, size);
CLUtilities.checkCLError(errorCodeBuffer);
return buffer;
}
/**
* Creates an opencl buffer based on a cpu buffer
* @param flags The flags for the buffer
* @param hostBuffer The cpu buffer to base the opencl buffer on
* @return The opencl buffer
*/
public CLBuffer createBuffer(long flags, IntBuffer hostBuffer){
CLBuffer buffer = new CLBuffer(this, flags, hostBuffer);
CLUtilities.checkCLError(errorCodeBuffer);
return buffer;
}
/**
* Creates a 2d opencl image buffer
* @param flags The flags for creation of the image
* @param width The width of the image
* @param height The height of the image
* @return The opencl image wrapper object
*/
public CLImage createImage(long flags, long width, long height){
CLImage image = new CLImage(this, flags, width, height);
CLUtilities.checkCLError(errorCodeBuffer);
return image;
}
/**
* Creates a 3d opencl image buffer
* @param flags The flags for creation of the image
* @param width The width of the image
* @param height The height of the image
* @param depth The depth of the image
* @return The opencl image wrapper object
*/
public CLImage createImage(long flags, long width, long height, long depth){
CLImage image = new CLImage(this, flags, width, height, depth);
CLUtilities.checkCLError(errorCodeBuffer);
return image;
}
/**
* Creates a 3d opencl image buffer
* @param flags The flags for creation of the image
* @param width The width of the image
* @param height The height of the image
* @param depth The depth of the image
* @param data The data to init the image with
* @return The opencl image wrapper object
*/
public CLImage createImage(long flags, long width, long height, long depth, ByteBuffer data){
CLImage image = new CLImage(this, flags, width, height, depth, data);
CLUtilities.checkCLError(errorCodeBuffer);
return image;
}
/**
* Creates a command queue for this context
* @return The command queue
*/
public CLCommandQueue createCommandQueue(LWJGLContext lwjglContext){
return new CLCommandQueue(lwjglContext, this);
}
/**
* Gets the opencl device this context was created on
* @return The device
*/
public CLDevice getDevice(){
return device;
}
/**
* Creates an opencl program with a provided source
* @param source The source
* @return The program
*/
public CLProgram createProgram(String source){
CLProgram program = CLProgram.createProgram(this, source);
return program;
}
/**
* Creates an opencl kernel from the specified program
* @param program The program
* @param kernelName The name of the kernel in the program source
* @return The kernel object
*/
public CLKernel createKernel(CLProgram program, String kernelName){
CLKernel kernel = new CLKernel(this, program, kernelName);
return kernel;
}
}

68
src/main/java/electrosphere/opencl/CLDevice.java
View File

@ -1,68 +0,0 @@
package electrosphere.opencl;
import org.lwjgl.opencl.CL30;
/**
* Wrapper object for an opencl device
*/
public class CLDevice {
//the id of the device
long id;
//the name of the device
String deviceName;
//true if the device is a gpu
boolean isGpu = false;
//true if the device is available
boolean isAvailable = false;
/**
* Creates a CL device wrapper object
* @param id the id of the device
* @param isGpu true if the device is a gpu
*/
protected CLDevice(long id){
this.id = id;
//get device name
deviceName = CLUtilities.getDeviceInfoStringUTF8(id, CL30.CL_DEVICE_NAME);
//get if device is gpu
long deviceType = CLUtilities.getDeviceInfoLong(id, CL30.CL_DEVICE_TYPE);
if (deviceType == CL30.CL_DEVICE_TYPE_GPU) {
isGpu = true;
}
//check if is available
long isAvailableRaw = CLUtilities.getDeviceInfoLong(id, CL30.CL_DEVICE_AVAILABLE);
if(isAvailableRaw == CL30.CL_TRUE){
isAvailable = true;
}
}
/**
* Gets the id of the device
* @return the id of the device
*/
public long getId(){
return id;
}
/**
* True if the device is a gpu
* @return True if the device is a gpu
*/
public boolean isGpu(){
return isGpu;
}
/**
* Gets the name of the device
* @return The name of the device
*/
public String getDeviceName(){
return deviceName;
}
}

65
src/main/java/electrosphere/opencl/CLEvent.java
View File

@ -1,65 +0,0 @@
package electrosphere.opencl;
import java.nio.ByteBuffer;
import org.lwjgl.BufferUtils;
import org.lwjgl.PointerBuffer;
import org.lwjgl.opencl.CL30;
import org.lwjgl.system.MemoryStack;
import org.lwjgl.system.MemoryUtil;
/**
* Wrapper for an opencl event
*/
public class CLEvent {
//The pointer for the event
long pointer;
/**
* Constructs a wrapper object around an opencl event
* @param pointer The pointer for the event
*/
protected CLEvent(long pointer){
this.pointer = pointer;
}
/**
* Gets the pointer for this event
* @return The pointer as a long
*/
public long getPointer(){
return pointer;
}
/**
* Queries the status of the event from opencl
* @return The event status code
*/
public int getStatus(){
int rVal = 0;
try(MemoryStack stack = MemoryStack.stackPush()){
//create buffers for call
ByteBuffer returnValueBuffer = stack.malloc(1);
PointerBuffer sizeBuffer = stack.mallocPointer(1);
//submit request to opencl
int errorCode = CL30.clGetEventInfo(this.getPointer(),CL30.CL_EVENT_COMMAND_EXECUTION_STATUS,returnValueBuffer,sizeBuffer);
//check error
CLUtilities.checkCLError(errorCode);
//get return value
rVal = returnValueBuffer.getInt();
// //free buffers
// MemoryUtil.memFree(returnValueBuffer);
// sizeBuffer.free();
}
//return
return rVal;
}
}

218
src/main/java/electrosphere/opencl/CLImage.java
View File

@ -1,218 +0,0 @@
package electrosphere.opencl;
import java.nio.ByteBuffer;
import java.nio.IntBuffer;
import org.lwjgl.BufferUtils;
import org.lwjgl.opencl.CL30;
import org.lwjgl.opencl.CLImageDesc;
import org.lwjgl.opencl.CLImageFormat;
import org.lwjgl.system.MemoryUtil;
/**
* Wrapper object for an opencl image
*/
public class CLImage {
//the pointer to the image
long pointer;
//The dimensions of the image
long width;
long height;
long depth;
//true if device can read from the image
boolean canRead;
//true if device can write to the image
boolean canWrite;
//the number of bytes per pixel. This is used by the command queue fetching tasks to allocate buffer sizes
long byteWidth;
/**
* Creates an empty 2d image
* @param context The context to create the image in
* @param flags The flags for the image
* @param width The width of the image
* @param height The height of the image
*/
protected CLImage(CLContext context, long flags, long width, long height){
//set initial values
this.width = width;
this.height = height;
this.depth = 0;
//parse flags
this.parseFlags(flags);
//
//Get the image format
CLImageFormat imageFormat = CLImageFormat.create();
imageFormat.image_channel_data_type(CL30.CL_SNORM_INT8);
byteWidth = 4;
imageFormat.image_channel_order(CL30.CL_RGBA);
//
//Get the image description
CLImageDesc imageDesc = CLImageDesc.create();
imageDesc.image_type(CL30.CL_MEM_OBJECT_IMAGE2D);
imageDesc.image_width(width);
imageDesc.image_height(height);
imageDesc.image_depth(0); //only used for 3d image
imageDesc.image_array_size(1); //only used for 1d and 2d images
imageDesc.image_row_pitch(0); //must be 0 if host ptr is null
imageDesc.image_slice_pitch(0); //must be 0 if host ptr is null
//
//host pointer (the data to prepopulate with)
IntBuffer hostPointerBuffer = null; //BufferUtils.createIntBuffer(1);
//
//Create the image object
pointer = CL30.clCreateImage(context.getPointer(), flags, imageFormat, imageDesc, hostPointerBuffer, context.getErrorCodeBuffer());
//error check
CLUtilities.checkCLError(context.getErrorCodeBuffer());
}
/**
* Creates an empty 3d image
* @param context The context to create the image in
* @param flags The flags for the image
* @param width The width of the image
* @param height The height of the image
* @param depth The depth of the image
*/
protected CLImage(CLContext context, long flags, long width, long height, long depth){
//set initial values
this.width = width;
this.height = height;
this.depth = depth;
//parse flags
this.parseFlags(flags);
//
//Get the image format
CLImageFormat imageFormat = CLImageFormat.create();
imageFormat.image_channel_data_type(CL30.CL_SNORM_INT8);
byteWidth = 4;
imageFormat.image_channel_order(CL30.CL_RGBA);
//
//Get the image description
CLImageDesc imageDesc = CLImageDesc.create();
imageDesc.image_type(CL30.CL_MEM_OBJECT_IMAGE3D);
imageDesc.image_width(width);
imageDesc.image_height(height);
imageDesc.image_depth(depth); //only used for 3d image
imageDesc.image_array_size(1); //only used for 1d and 2d images
imageDesc.image_row_pitch(0); //must be 0 if host ptr is null
imageDesc.image_slice_pitch(0); //must be 0 if host ptr is null
//
//host pointer (the data to prepopulate with)
IntBuffer hostPointerBuffer = null; //BufferUtils.createIntBuffer(1);
//
//Create the image object
pointer = CL30.clCreateImage(context.getPointer(), flags, imageFormat, imageDesc, hostPointerBuffer, context.getErrorCodeBuffer());
}
/**
* Creates an empty 3d image
* @param context The context to create the image in
* @param flags The flags for the image
* @param width The width of the image
* @param height The height of the image
* @param depth The depth of the image
*/
protected CLImage(CLContext context, long flags, long width, long height, long depth, ByteBuffer data){
//set initial values
this.width = width;
this.height = height;
this.depth = depth;
//parse flags
this.parseFlags(flags);
//
//Get the image format
CLImageFormat imageFormat = CLImageFormat.create();
imageFormat.image_channel_data_type(CL30.CL_SNORM_INT8);
byteWidth = 4;
imageFormat.image_channel_order(CL30.CL_RGBA);
//
//Get the image description
CLImageDesc imageDesc = CLImageDesc.create();
imageDesc.image_type(CL30.CL_MEM_OBJECT_IMAGE3D);
imageDesc.image_width(width);
imageDesc.image_height(height);
imageDesc.image_depth(depth); //only used for 3d image
imageDesc.image_array_size(1); //only used for 1d and 2d images
imageDesc.image_row_pitch(width * byteWidth); //must be 0 if host ptr is null
imageDesc.image_slice_pitch(width * height * byteWidth); //must be 0 if host ptr is null
//
//Create the image object
pointer = CL30.clCreateImage(context.getPointer(), flags, imageFormat, imageDesc, data, context.getErrorCodeBuffer());
CLUtilities.checkCLError(context);
}
/**
* Parses the flags passed in on image creation to set variables (eg read/write status)
* @param flags The flags passed in
*/
private void parseFlags(long flags){
if((CL30.CL_MEM_READ_ONLY & flags) != 0){
this.canRead = true;
}
if((CL30.CL_MEM_WRITE_ONLY & flags) != 0){
this.canWrite = true;
}
}
/**
* Gets the pointer for the image
* @return The image pointer
*/
public long getPointer(){
return pointer;
}
/**
* Gets the width of the image
* @return The width of the image
*/
public long getWidth(){
return width;
}
/**
* Gets the height of the image
* @return The height of the image
*/
public long getHeight(){
return height;
}
/**
* Gets the depth of the image
* @return The depth of the image
*/
public long getDepth(){
return depth;
}
/**
* Gets the byte width of the image
* @return The byte width
*/
public long getByteWidth(){
return byteWidth;
}
}

32
src/main/java/electrosphere/opencl/CLKernel.java
View File

@ -1,32 +0,0 @@
package electrosphere.opencl;
import org.lwjgl.opencl.CL30;
/**
* A wrapper object for an opencl kernel
*/
public class CLKernel {
//The pointer to the kernel
long pointer;
/**
* Creates a cl kernel from a context and a program
* @param context The opencl context
* @param program The opencl program
* @param kernelName The name of the kernel
*/
protected CLKernel(CLContext context, CLProgram program, String kernelName){
pointer = CL30.clCreateKernel(program.getPointer(), kernelName, context.getErrorCodeBuffer());
CLUtilities.checkCLError(context.getErrorCodeBuffer());
}
/**
* Gets the pointer for this kernel
* @return The pointer as a long
*/
public long getPointer(){
return pointer;
}
}

153
src/main/java/electrosphere/opencl/CLPlatform.java
View File

@ -1,153 +0,0 @@
package electrosphere.opencl;
import java.nio.ByteBuffer;
import java.nio.CharBuffer;
import java.nio.IntBuffer;
import java.nio.LongBuffer;
import java.util.ArrayList;
import java.util.Collections;
import java.util.HashSet;
import java.util.LinkedList;
import java.util.List;
import java.util.Set;
import org.lwjgl.PointerBuffer;
import org.lwjgl.opencl.CL;
import org.lwjgl.opencl.CL30;
import org.lwjgl.opencl.CLCapabilities;
import org.lwjgl.system.MemoryStack;
/**
* An object that encapsulates an openCL platform
*/
public class CLPlatform {
//the id of the platform
long id;
//the capabilities of the platform
CLCapabilities caps;
//Information about the platform
String vendor;
String name;
String version;
// the list of all cl devices on this platform
List<CLDevice> clDevices;
/**
* Creates a platform object
* @param id The id of the platform
*/
protected CLPlatform(long id, CLCapabilities caps, String vendor, String name, String version, List<CLDevice> clDevices){
this.id = id;
this.caps = caps;
this.vendor = vendor;
this.name = name;
this.version = version;
this.clDevices = clDevices;
}
/**
* Gets a set containing all currently available platforms
* @return The set containing all currently available platforms
*/
public static Set<CLPlatform> getPlatforms(){
Set<CLPlatform> platforms = new HashSet<CLPlatform>();
try(MemoryStack stack = MemoryStack.stackPush()){
//check if there are any platforms
IntBuffer pi = stack.mallocInt(1);
CLUtilities.checkCLError(CL30.clGetPlatformIDs(null, pi));
if (pi.get(0) == 0) {
throw new IllegalStateException("No OpenCL platforms found.");
}
//if there are platforms, get all platform IDs
PointerBuffer platformIDs = stack.mallocPointer(pi.get(0));
CLUtilities.checkCLError(CL30.clGetPlatformIDs(platformIDs, (IntBuffer)null));
//add all platforms to the returned list
for (int i = 0; i < platformIDs.capacity(); i++) {
//get basic identification
long platformId = platformIDs.get(i);
CLCapabilities caps = CL.createPlatformCapabilities(platformId);
String vendor = CLUtilities.getPlatformInfoStringUTF8(platformId, CL30.CL_PLATFORM_VENDOR);
String name = CLUtilities.getPlatformInfoStringUTF8(platformId, CL30.CL_PLATFORM_NAME);
String version = CLUtilities.getPlatformInfoStringUTF8(platformId, CL30.CL_PLATFORM_VERSION);
//get devices on the platform
int errcode = CL30.clGetDeviceIDs(platformId, CL30.CL_DEVICE_TYPE_GPU, null, pi);
List<CLDevice> devices = new LinkedList<CLDevice>();
if(errcode != CL30.CL_DEVICE_NOT_FOUND){
CLUtilities.checkCLError(errcode);
PointerBuffer deviceIDs = stack.mallocPointer(pi.get(0));
CLUtilities.checkCLError(CL30.clGetDeviceIDs(platformId, CL30.CL_DEVICE_TYPE_ALL, deviceIDs, (IntBuffer)null));
for (int j = 0; j < deviceIDs.capacity(); j++) {
long deviceId = deviceIDs.get(j);
devices.add(new CLDevice(deviceId));
}
}
platforms.add(new CLPlatform(platformId, caps, vendor, name, version, devices));
}
if (platforms.isEmpty()) {
throw new IllegalStateException("No OpenCL platform found.");
}
}
return platforms;
}
/**
* Gets the id of the platform
* @return the id of the platfomr
*/
public long getId(){
return id;
}
/**
* Gets the capabilities object of the platform
* @return The capabilities
*/
public CLCapabilities getCapabilities(){
return caps;
}
/**
* Gets the name of the vendor the platform
* @return The name of the vendor of the platform
*/
public String getVendor(){
return vendor;
}
/**
* Gets the name of the platform
* @return The name of the platform
*/
public String getName(){
return name;
}
/**
* Gets the version of opencl on the platform
* @return The version of opencl on the platform
*/
public String getVersion(){
return name;
}
/**
* Gets the list of cl devices on this platform
* @return The list of devices
*/
public List<CLDevice> getCLDevices(){
return clDevices;
}
}

70
src/main/java/electrosphere/opencl/CLProgram.java
View File

@ -1,70 +0,0 @@
package electrosphere.opencl;
import java.util.concurrent.CountDownLatch;
import org.lwjgl.opencl.CL30;
import org.lwjgl.opencl.CLProgramCallback;
import org.lwjgl.system.MemoryUtil;
/**
* Wrapper object for an opencl program
*/
public class CLProgram {
//the pointer to the program
long pointer;
/**
* Creates an opencl program from a context and a source
* @param context The context
* @param source The source
* @return The program
*/
protected static CLProgram createProgram(CLContext context, String source){
CLProgram program = new CLProgram();
//create object
program.pointer = CL30.clCreateProgramWithSource(context.getPointer(), source, context.getErrorCodeBuffer());
CLUtilities.checkCLError(context.getErrorCodeBuffer());
//latch for blocking execution until the program finishes compiling
CountDownLatch latch = new CountDownLatch(1);
//build the program
String options = "";
CLProgramCallback buildCallback;
int errcode = CL30.clBuildProgram(program.pointer, context.getDevice().getId(), options, buildCallback = CLProgramCallback.create((programId, user_data) -> {
System.out.println(String.format(
"The cl_program [0x%X] was built %s",
programId,
CLUtilities.getProgramBuildInfoInt(programId, context.getDevice().getId(), CL30.CL_PROGRAM_BUILD_STATUS) == CL30.CL_SUCCESS ? "successfully" : "unsuccessfully"
));
String log = CLUtilities.getProgramBuildInfoStringASCII(programId, context.getDevice().getId(), CL30.CL_PROGRAM_BUILD_LOG);
if (!log.isEmpty()) {
System.out.println(String.format("BUILD LOG:\n----\n%s\n-----", log));
}
latch.countDown();
}), MemoryUtil.NULL);
CLUtilities.checkCLError(errcode);
//Block execution until the program finishes building
try {
latch.await();
} catch (InterruptedException e) {
throw new RuntimeException(e);
}
buildCallback.free();
return program;
}
/**
* Gets the pointer for this program object
* @return The pointer as a long
*/
public long getPointer(){
return pointer;
}
}

41
src/main/java/electrosphere/opencl/CLReadImageResult.java
View File

@ -1,41 +0,0 @@
package electrosphere.opencl;
import java.nio.ByteBuffer;
/**
* A pair of an opencl event and the buffer from the read operation. Used so we can pass back both from the call.
*/
public class CLReadImageResult {
//the buffer
ByteBuffer buffer;
//the event
CLEvent event;
/**
* Constructs the pair
* @param buffer The buffer
* @param event The event
*/
protected CLReadImageResult(ByteBuffer buffer, CLEvent event){
this.buffer = buffer;
this.event = event;
}
/**
* Gets the buffer
* @return The buffer
*/
public ByteBuffer getBuffer(){
return buffer;
}
/**
* Gets the event
* @return The event
*/
public CLEvent getEvent(){
return event;
}
}

181
src/main/java/electrosphere/opencl/CLUtilities.java
View File

@ -1,181 +0,0 @@
package electrosphere.opencl;
import java.nio.ByteBuffer;
import java.nio.IntBuffer;
import java.nio.LongBuffer;
import org.lwjgl.BufferUtils;
import org.lwjgl.PointerBuffer;
import org.lwjgl.opencl.CL30;
import org.lwjgl.system.MemoryStack;
import org.lwjgl.system.MemoryUtil;
import static org.lwjgl.opencl.CL30.*;
import static org.lwjgl.system.MemoryStack.*;
import static org.lwjgl.system.MemoryUtil.*;
/**
* Utilities for all opencl files (for instance error checking functions)
*/
public class CLUtilities {
/**
* Checks an opencl error code
* @param errcode The error code
*/
public static void checkCLError(IntBuffer errcode) {
checkCLError(errcode.get(errcode.position()));
}
/**
* Checks an opencl error code
* @param errcode The error code
*/
public static void checkCLError(int errcode) {
if (errcode != CL_SUCCESS) {
throw new RuntimeException(String.format("OpenCL error [%d]", errcode));
}
}
/**
* Checks an opencl error code
* @param context The context containing the error code buffer
*/
public static void checkCLError(CLContext context) {
int code = context.getErrorCodeBuffer().get();
if (code != CL_SUCCESS) {
throw new RuntimeException(String.format("OpenCL error [%d]", code));
}
context.getErrorCodeBuffer().flip();
}
/**
* Gets a platform info as a utf string
* @param platformId The platform id
* @param paramName The parameter name
* @return The string containing the information
*/
public static String getPlatformInfoStringUTF8(long platformId, int paramName) {
String rVal = "";
try(MemoryStack stack = MemoryStack.stackPush()){
PointerBuffer resultSizeBuffer = stack.mallocPointer(1);
checkCLError(CL30.clGetPlatformInfo(platformId, paramName, (ByteBuffer)null, resultSizeBuffer));
int bytes = (int)resultSizeBuffer.get(0);
ByteBuffer buffer = stack.malloc(bytes);
checkCLError(CL30.clGetPlatformInfo(platformId, paramName, buffer, null));
rVal = memUTF8(buffer, bytes - 1);
}
return rVal;
}
/**
* Gets an opencl device parameter as a utf8 string
* @param deviceId The device id
* @param paramName The parameter name
* @return The string
*/
public static String getDeviceInfoStringUTF8(long deviceId, int paramName) {
String rVal = "";
try(MemoryStack stack = MemoryStack.stackPush()){
PointerBuffer pp = stack.mallocPointer(1);
checkCLError(clGetDeviceInfo(deviceId, paramName, (ByteBuffer)null, pp));
int bytes = (int)pp.get(0);
ByteBuffer buffer = stack.malloc(bytes);
checkCLError(clGetDeviceInfo(deviceId, paramName, buffer, null));
rVal = memUTF8(buffer, bytes - 1);
}
return rVal;
}
/**
* Gets a device parameter's value as an int
* @param deviceId The id of the device
* @param paramName The parameter name
* @return The int
*/
public static int getDeviceInfoInt(long deviceId, int paramName) {
int rVal = 0;
try(MemoryStack stack = MemoryStack.stackPush()){
IntBuffer pl = stack.mallocInt(1);
checkCLError(clGetDeviceInfo(deviceId, paramName, pl, null));
rVal = pl.get(0);
}
return rVal;
}
/**
* Gets a device's parameter's value as a long
* @param deviceId The device's id
* @param paramName The parameter name
* @return The value as a long
*/
public static long getDeviceInfoLong(long deviceId, int paramName) {
long rVal = 0;
try(MemoryStack stack = MemoryStack.stackPush()){
LongBuffer pl = stack.mallocLong(1);
checkCLError(clGetDeviceInfo(deviceId, paramName, pl, null));
rVal = pl.get(0);
}
return rVal;
}
/**
* Gets a device's parameter's value as a boolean
* @param deviceId The device's id
* @param paramName The parameter name
* @return The value as a boolean
*/
public static boolean getDeviceInfoBoolean(long deviceId, int paramName) {
byte rVal = 0;
try(MemoryStack stack = MemoryStack.stackPush()){
ByteBuffer pl = stack.malloc(1);
checkCLError(clGetDeviceInfo(deviceId, paramName, pl, null));
rVal = pl.get(0);
}
return rVal == 0;
}
/**
* Gets an opencl program build parameter as an int
* @param programId The program id
* @param deviceId The device id
* @param parameterName The parameter name
* @return THe value of the parameter as an int
*/
public static int getProgramBuildInfoInt(long programId, long deviceId, int parameterName) {
int rVal = 0;
try(MemoryStack stack = MemoryStack.stackPush()){
IntBuffer pl = stack.mallocInt(1);
checkCLError(clGetProgramBuildInfo(programId, deviceId, parameterName, pl, null));
rVal = pl.get(0);
}
return rVal;
}
/**
* Gets an opencl program build parameter as a string
* @param programId The program id
* @param deviceId The device id
* @param parameterName The parameter name
* @return The value of the parameter as a string
*/
public static String getProgramBuildInfoStringASCII(long programId, long deviceId, int parameterName) {
String rVal = "";
try(MemoryStack stack = MemoryStack.stackPush()){
PointerBuffer pp = stack.mallocPointer(1);
checkCLError(clGetProgramBuildInfo(programId, deviceId, parameterName, (ByteBuffer)null, pp));
int bytes = (int)pp.get(0);
ByteBuffer buffer = stack.malloc(bytes);
checkCLError(clGetProgramBuildInfo(programId, deviceId, parameterName, buffer, null));
rVal = memASCII(buffer, bytes - 1);;
}
return rVal;
}
}

28
src/main/java/electrosphere/opencl/LWJGLContext.java
View File

@ -1,28 +0,0 @@
package electrosphere.opencl;
import org.lwjgl.system.MemoryStack;
/**
* A context for the lwjgl framework. Contains stack used by opencl api.
*/
public class LWJGLContext {
//The stack for memory allocation
MemoryStack stack;
/**
* Constructor
*/
public LWJGLContext(){
stack = MemoryStack.create(1000 * 1000 * 4 * 4);
}
/**
* Gets the memory stack for the current context
* @return The lwjgl context's memory stack
*/
public MemoryStack getStack(){
return stack;
}
}

21
src/main/java/electrosphere/render/GLFWContext.java
View File

@ -39,14 +39,19 @@ public class GLFWContext {
//Maximize it //Maximize it
GLFW.glfwMaximizeWindow(window); GLFW.glfwMaximizeWindow(window);
//grab actual framebuffer //grab actual framebuffer
IntBuffer xBuffer = BufferUtils.createIntBuffer(1); int bufferWidth = 0;
IntBuffer yBuffer = BufferUtils.createIntBuffer(1); int bufferHeight = 0;
GLFW.glfwGetFramebufferSize(window, xBuffer, yBuffer);
MemoryUtil.memFree(xBuffer); try(MemoryStack stack = MemoryStack.stackPush()){
MemoryUtil.memFree(yBuffer); IntBuffer xBuffer = MemoryUtil.memAllocInt(1);
IntBuffer yBuffer = MemoryUtil.memAllocInt(1);
int bufferWidth = xBuffer.get(); GLFW.glfwGetFramebufferSize(window, xBuffer, yBuffer);
int bufferHeight = yBuffer.get(); MemoryUtil.memFree(xBuffer);
MemoryUtil.memFree(yBuffer);
bufferWidth = xBuffer.get();
bufferHeight = yBuffer.get();
}
// //
// Attack controls callbacks // Attack controls callbacks

19
src/main/java/electrosphere/render/GridCell.java Normal file
View File

@ -0,0 +1,19 @@
package electrosphere.render;
import org.joml.Vector3f;
public class GridCell {
Vector3f[] points = new Vector3f[8]; //array of size 8
double[] val = new double[8]; //array of size 8
public void setValues(
Vector3f p1, Vector3f p2, Vector3f p3, Vector3f p4,
Vector3f p5, Vector3f p6, Vector3f p7, Vector3f p8,
double val1, double val2, double val3, double val4,
double val5, double val6, double val7, double val8
){
points[0] = p1; points[1] = p2; points[2] = p3; points[3] = p4;
points[4] = p5; points[5] = p6; points[6] = p7; points[7] = p8;
val[0] = val1; val[1] = val2; val[2] = val3; val[3] = val4;
val[4] = val5; val[5] = val6; val[6] = val7; val[7] = val8;
}
}

164
src/main/java/electrosphere/render/Mesh.java
View File

@ -38,7 +38,7 @@ public class Mesh {
static Matrix4f projectionMatrix = new Matrix4f(); static Matrix4f projectionMatrix = new Matrix4f();
static Matrix4f model = new Matrix4f().identity(); static Matrix4f model = new Matrix4f().identity();
static MemoryStack stack = MemoryStack.create(16 * 1000 * 1000); static MemoryStack stack = MemoryStack.create(32 * 1000 * 1000);
public static void meshInitially(FluidSim sim){ public static void meshInitially(FluidSim sim){
//create and bind vao //create and bind vao
@ -55,7 +55,7 @@ public class Mesh {
GL44.glBindBuffer(GL44.GL_ARRAY_BUFFER, vertBufferPtr); GL44.glBindBuffer(GL44.GL_ARRAY_BUFFER, vertBufferPtr);
try { try {
int vertexCount = data.vertices.size(); int vertexCount = data.vertices.size();
FloatBuffer vertData = BufferUtils.createFloatBuffer(vertexCount); FloatBuffer vertData = MemoryUtil.memAllocFloat(vertexCount);
for(float vertValue : data.vertices){ for(float vertValue : data.vertices){
vertData.put(vertValue); vertData.put(vertValue);
} }
@ -63,7 +63,7 @@ public class Mesh {
GL44.glBufferData(GL44.GL_ARRAY_BUFFER,vertData,GL44.GL_STATIC_DRAW); GL44.glBufferData(GL44.GL_ARRAY_BUFFER,vertData,GL44.GL_STATIC_DRAW);
GL44.glVertexAttribPointer(0, 3, GL44.GL_FLOAT, false, 0, 0); GL44.glVertexAttribPointer(0, 3, GL44.GL_FLOAT, false, 0, 0);
GL44.glEnableVertexAttribArray(0); GL44.glEnableVertexAttribArray(0);
// MemoryUtil.memFree(vertData); MemoryUtil.memFree(vertData);
} catch (NullPointerException ex){ } catch (NullPointerException ex){
ex.printStackTrace(); ex.printStackTrace();
} }
@ -77,13 +77,13 @@ public class Mesh {
GL44.glBindBuffer(GL44.GL_ELEMENT_ARRAY_BUFFER, faceBufferPtr); GL44.glBindBuffer(GL44.GL_ELEMENT_ARRAY_BUFFER, faceBufferPtr);
elementCount = data.elements.size(); elementCount = data.elements.size();
try { try {
IntBuffer elementArrayBufferData = BufferUtils.createIntBuffer(elementCount); IntBuffer elementArrayBufferData = MemoryUtil.memAllocInt(elementCount);
for(int element : data.elements){ for(int element : data.elements){
elementArrayBufferData.put(element); elementArrayBufferData.put(element);
} }
elementArrayBufferData.flip(); elementArrayBufferData.flip();
GL44.glBufferData(GL45.GL_ELEMENT_ARRAY_BUFFER,elementArrayBufferData,GL44.GL_STATIC_DRAW); GL44.glBufferData(GL45.GL_ELEMENT_ARRAY_BUFFER,elementArrayBufferData,GL44.GL_STATIC_DRAW);
// MemoryUtil.memFree(elementArrayBufferData); MemoryUtil.memFree(elementArrayBufferData);
} catch (NullPointerException ex){ } catch (NullPointerException ex){
ex.printStackTrace(); ex.printStackTrace();
} }
@ -100,7 +100,7 @@ public class Mesh {
int normalCount = data.normals.size() / 3; int normalCount = data.normals.size() / 3;
FloatBuffer NormalArrayBufferData; FloatBuffer NormalArrayBufferData;
if(normalCount > 0){ if(normalCount > 0){
NormalArrayBufferData = BufferUtils.createFloatBuffer(normalCount * 3); NormalArrayBufferData = MemoryUtil.memAllocFloat(normalCount * 3);
float[] temp = new float[3]; float[] temp = new float[3];
for(float normalValue : data.normals){ for(float normalValue : data.normals){
NormalArrayBufferData.put(normalValue); NormalArrayBufferData.put(normalValue);
@ -202,67 +202,66 @@ public class Mesh {
//generate verts //generate verts
TerrainChunkData data = generateTerrainChunkData(sim.getData()); TerrainChunkData data = generateTerrainChunkData(sim.getData());
try(MemoryStack stackLocal = stack.push()){
// //
//Buffer data to GPU //Buffer data to GPU
// //
GL44.glBindBuffer(GL44.GL_ARRAY_BUFFER, vertBufferPtr); GL44.glBindBuffer(GL44.GL_ARRAY_BUFFER, vertBufferPtr);
try { try {
int vertexCount = data.vertices.size(); int vertexCount = data.vertices.size();
FloatBuffer vertData = stackLocal.mallocFloat(vertexCount); FloatBuffer vertData = MemoryUtil.memAllocFloat(vertexCount);
for(float vertValue : data.vertices){ for(float vertValue : data.vertices){
vertData.put(vertValue); vertData.put(vertValue);
}
vertData.flip();
GL44.glBufferData(GL44.GL_ARRAY_BUFFER,vertData,GL44.GL_STATIC_DRAW);
GL44.glVertexAttribPointer(0, 3, GL44.GL_FLOAT, false, 0, 0);
GL44.glEnableVertexAttribArray(0);
} catch (NullPointerException ex){
ex.printStackTrace();
} }
vertData.flip();
GL44.glBufferData(GL44.GL_ARRAY_BUFFER,vertData,GL44.GL_STATIC_DRAW);
GL44.glVertexAttribPointer(0, 3, GL44.GL_FLOAT, false, 0, 0);
GL44.glEnableVertexAttribArray(0);
MemoryUtil.memFree(vertData);
} catch (NullPointerException ex){
ex.printStackTrace();
}
//
// // FACES
// FACES //
// GL44.glBindBuffer(GL44.GL_ELEMENT_ARRAY_BUFFER, faceBufferPtr);
GL44.glBindBuffer(GL44.GL_ELEMENT_ARRAY_BUFFER, faceBufferPtr); elementCount = data.elements.size();
elementCount = data.elements.size(); try {
try { IntBuffer elementArrayBufferData = MemoryUtil.memAllocInt(elementCount);
IntBuffer elementArrayBufferData = stackLocal.mallocInt(elementCount); for(int element : data.elements){
for(int element : data.elements){ elementArrayBufferData.put(element);
elementArrayBufferData.put(element);
}
elementArrayBufferData.flip();
GL44.glBufferData(GL45.GL_ELEMENT_ARRAY_BUFFER,elementArrayBufferData,GL44.GL_STATIC_DRAW);
} catch (NullPointerException ex){
ex.printStackTrace();
} }
elementArrayBufferData.flip();
GL44.glBufferData(GL45.GL_ELEMENT_ARRAY_BUFFER,elementArrayBufferData,GL44.GL_STATIC_DRAW);
MemoryUtil.memFree(elementArrayBufferData);
} catch (NullPointerException ex){
// ex.printStackTrace();
// NORMALS }
//
GL44.glBindBuffer(GL44.GL_ARRAY_BUFFER, normBufferPtr);
try {
int normalCount = data.normals.size() / 3; //
FloatBuffer NormalArrayBufferData; // NORMALS
if(normalCount > 0){ //
NormalArrayBufferData = stackLocal.mallocFloat(normalCount * 3); GL44.glBindBuffer(GL44.GL_ARRAY_BUFFER, normBufferPtr);
for(float normalValue : data.normals){ try {
NormalArrayBufferData.put(normalValue); int normalCount = data.normals.size() / 3;
} FloatBuffer NormalArrayBufferData;
NormalArrayBufferData.flip(); if(normalCount > 0){
GL44.glBufferData(GL44.GL_ARRAY_BUFFER,NormalArrayBufferData,GL44.GL_STATIC_DRAW); NormalArrayBufferData = MemoryUtil.memAllocFloat(normalCount * 3);
GL44.glVertexAttribPointer(1, 3, GL44.GL_FLOAT, false, 0, 0); for(float normalValue : data.normals){
GL44.glEnableVertexAttribArray(1); NormalArrayBufferData.put(normalValue);
} }
} catch (NullPointerException ex){ NormalArrayBufferData.flip();
ex.printStackTrace(); GL44.glBufferData(GL44.GL_ARRAY_BUFFER,NormalArrayBufferData,GL44.GL_STATIC_DRAW);
GL44.glVertexAttribPointer(1, 3, GL44.GL_FLOAT, false, 0, 0);
GL44.glEnableVertexAttribArray(1);
MemoryUtil.memFree(NormalArrayBufferData);
} }
} catch (NullPointerException ex){
ex.printStackTrace();
} }
} }
@ -613,33 +612,6 @@ public class Mesh {
(byte)0x13 (byte)0x13
}; };
static class Triangle {
int[] indices = new int[3]; //array of size 3
public Triangle(int index0, int index1, int index2){
indices[0] = index0;
indices[1] = index1;
indices[2] = index2;
}
}
static class GridCell {
Vector3f[] points = new Vector3f[8]; //array of size 8
double[] val = new double[8]; //array of size 8
public void setValues(
Vector3f p1, Vector3f p2, Vector3f p3, Vector3f p4,
Vector3f p5, Vector3f p6, Vector3f p7, Vector3f p8,
double val1, double val2, double val3, double val4,
double val5, double val6, double val7, double val8
){
points[0] = p1; points[1] = p2; points[2] = p3; points[3] = p4;
points[4] = p5; points[5] = p6; points[6] = p7; points[7] = p8;
val[0] = val1; val[1] = val2; val[2] = val3; val[3] = val4;
val[4] = val5; val[5] = val6; val[6] = val7; val[7] = val8;
}
}
protected static int polygonize( protected static int polygonize(
@ -897,20 +869,6 @@ public class Mesh {
return rVal; return rVal;
} }
static class TerrainChunkData {
List<Float> vertices;
List<Float> normals;
List<Integer> elements;
TerrainChunkData(List<Float> vertices, List<Float> normals, List<Integer> elements){
this.vertices = vertices;
this.normals = normals;
this.elements = elements;
}
}
//TODO: more optimal key creation //TODO: more optimal key creation
private static String getVertKeyFromPoints(float x, float y, float z){ private static String getVertKeyFromPoints(float x, float y, float z){
return x + "_" + y + "_" + z; return x + "_" + y + "_" + z;
@ -934,5 +892,7 @@ public class Mesh {
rVal = rVal.mul(proportion0).add(new Vector3f(normal1).mul(proportion1)); rVal = rVal.mul(proportion0).add(new Vector3f(normal1).mul(proportion1));
return rVal; return rVal;
} }
static TerrainChunkData data = generateTerrainChunkData(new float[FluidSim.DIM * FluidSim.DIM * FluidSim.DIM]);
} }

15
src/main/java/electrosphere/render/TerrainChunkData.java Normal file
View File

@ -0,0 +1,15 @@
package electrosphere.render;
import java.util.List;
public class TerrainChunkData {
List<Float> vertices;
List<Float> normals;
List<Integer> elements;
TerrainChunkData(List<Float> vertices, List<Float> normals, List<Integer> elements){
this.vertices = vertices;
this.normals = normals;
this.elements = elements;
}
}

11
src/main/java/electrosphere/render/Triangle.java Normal file
View File

@ -0,0 +1,11 @@
package electrosphere.render;
public class Triangle {
int[] indices = new int[3]; //array of size 3
public Triangle(int index0, int index1, int index2){
indices[0] = index0;
indices[1] = index1;
indices[2] = index2;
}
}

15
src/main/resources/shader.fs
View File

@ -1,7 +1,5 @@
#version 330 core #version 330 core
#define NR_POINT_LIGHTS 10
out vec4 FragColor; out vec4 FragColor;
@ -26,19 +24,7 @@ void main(){
vec3 norm = normalize(Normal); vec3 norm = normalize(Normal);
vec3 viewDir = normalize(viewPos - FragPos); vec3 viewDir = normalize(viewPos - FragPos);
// //grab light intensity
// float lightIntensity = calcLightIntensityTotal(norm);
// //get color of base texture
// vec3 textureColor = vec3(1);
// //calculate final color
// vec3 finalColor = textureColor * lightIntensity;
vec3 lightAmount = CalcDirLight(norm, viewDir); vec3 lightAmount = CalcDirLight(norm, viewDir);
// for(int i = 0; i < NR_POINT_LIGHTS; i++){
// lightAmount += CalcPointLight(i, norm, FragPos, viewDir);
// }
vec3 color = vec3(0.3,0.7,0.9) * lightAmount; vec3 color = vec3(0.3,0.7,0.9) * lightAmount;
@ -54,7 +40,6 @@ vec3 CalcDirLight(vec3 normal, vec3 viewDir){
vec3 reflectDir = reflect(-lightDir, normal); vec3 reflectDir = reflect(-lightDir, normal);
float spec = pow(max(dot(viewDir, reflectDir), 0.0), 0.6); float spec = pow(max(dot(viewDir, reflectDir), 0.0), 0.6);
// combine results // combine results
// vec3 texColor = texture(material.diffuse, TexCoord).rgb;
vec3 diffuse = dLDiffuse * diff; vec3 diffuse = dLDiffuse * diff;
vec3 specular = spec * color; vec3 specular = spec * color;

4
src/main/resources/shader.vs
View File

@ -1,10 +1,6 @@
//Vertex Shader //Vertex Shader
#version 330 core #version 330 core
//defines
#define TEXTURE_MAP_SCALE 3.0
//input buffers //input buffers
layout (location = 0) in vec3 aPos; layout (location = 0) in vec3 aPos;
layout (location = 1) in vec3 aNormal; layout (location = 1) in vec3 aNormal;