Skip to content
Snippets Groups Projects
Commit 6319738b authored by Ryan-LaMothe-PNNL's avatar Ryan-LaMothe-PNNL
Browse files

Merge pull request #21 from ekasitk/mdarray 

Resolve exception in kernel.dispose()
parents aaf0ae64 b0ba6623
No related branches found
No related tags found
No related merge requests found
Hide whitespace changes
Inline Side-by-side
Showing
with 236 additions and 186 deletions
com.amd.aparapi.jni/src/cpp/runKernel/Aparapi.cpp
+ 8
− 5
View file @ 6319738b
...@@ -422,10 +422,10 @@ void updateBuffer(JNIEnv* jenv, JNIContext* jniContext, KernelArg* arg, int& arg ...@@ -422,10 +422,10 @@ void updateBuffer(JNIEnv* jenv, JNIContext* jniContext, KernelArg* arg, int& arg
fprintf(stderr, "runKernel arg %d %s, length = %d\n", argIdx, arg->name, buffer->lens[i]); fprintf(stderr, "runKernel arg %d %s, length = %d\n", argIdx, arg->name, buffer->lens[i]);
} }
argPos++; argPos++;
status = clSetKernelArg(jniContext->kernel, argPos, sizeof(cl_uint), &(buffer->dims[i])); status = clSetKernelArg(jniContext->kernel, argPos, sizeof(cl_uint), &(buffer->offsets[i]));
if(status != CL_SUCCESS) throw CLException(status,"clSetKernelArg (buffer dimension)"); if(status != CL_SUCCESS) throw CLException(status,"clSetKernelArg (buffer offset)");
if (config->isVerbose()){ if (config->isVerbose()){
fprintf(stderr, "runKernel arg %d %s, dim = %d\n", argIdx, arg->name, buffer->dims[i]); fprintf(stderr, "runKernel arg %d %s, offsets = %d\n", argIdx, arg->name, buffer->offsets[i]);
} }
} }
} }
...@@ -469,7 +469,7 @@ void processArray(JNIEnv* jenv, JNIContext* jniContext, KernelArg* arg, int& arg ...@@ -469,7 +469,7 @@ void processArray(JNIEnv* jenv, JNIContext* jniContext, KernelArg* arg, int& arg
arg->pin(jenv); arg->pin(jenv);
if (config->isVerbose()) { if (config->isVerbose()) {
fprintf(stderr, "runKernel: arrayOrBuf ref %p, oldAddr=%p, newAddr=%p, ref.mem=%p isCopy=%s\n", fprintf(stderr, "runKernel: array ref %p, oldAddr=%p, newAddr=%p, ref.mem=%p isCopy=%s\n",
arg->arrayBuffer->javaArray, arg->arrayBuffer->javaArray,
prevAddr, prevAddr,
arg->arrayBuffer->addr, arg->arrayBuffer->addr,
...@@ -528,8 +528,11 @@ void processBuffer(JNIEnv* jenv, JNIContext* jniContext, KernelArg* arg, int& ar ...@@ -528,8 +528,11 @@ void processBuffer(JNIEnv* jenv, JNIContext* jniContext, KernelArg* arg, int& ar
arg->aparapiBuffer->write.valid = false; arg->aparapiBuffer->write.valid = false;
} }
// TODO: check if the object was moved and required re-flatten
arg->aparapiBuffer->flatten(jenv,arg);
if (config->isVerbose()) { if (config->isVerbose()) {
fprintf(stderr, "runKernel: arrayOrBuf addr=%p, ref.mem=%p\n", fprintf(stderr, "runKernel: Buf addr=%p, ref.mem=%p\n",
arg->aparapiBuffer->data, arg->aparapiBuffer->data,
arg->aparapiBuffer->mem); arg->aparapiBuffer->mem);
fprintf(stderr, "at memory addr %p, contents: ", arg->aparapiBuffer->data); fprintf(stderr, "at memory addr %p, contents: ", arg->aparapiBuffer->data);
......
com.amd.aparapi.jni/src/cpp/runKernel/AparapiBuffer.cpp
+ 126
− 107
View file @ 6319738b
...@@ -42,7 +42,7 @@ ...@@ -42,7 +42,7 @@
AparapiBuffer::AparapiBuffer(): AparapiBuffer::AparapiBuffer():
javaObject((jobject) 0), javaObject((jobject) 0),
numDims(0), numDims(0),
dims(NULL), offsets(NULL),
lengthInBytes(0), lengthInBytes(0),
mem((cl_mem) 0), mem((cl_mem) 0),
data(NULL), data(NULL),
...@@ -58,11 +58,11 @@ AparapiBuffer::AparapiBuffer(void* _data, cl_uint* _lens, cl_uint _numDims, long ...@@ -58,11 +58,11 @@ AparapiBuffer::AparapiBuffer(void* _data, cl_uint* _lens, cl_uint _numDims, long
mem((cl_mem) 0), mem((cl_mem) 0),
memMask((cl_uint)0) memMask((cl_uint)0)
{ {
dims = new cl_uint[_numDims]; offsets = new cl_uint[_numDims];
for(int i = 0; i < _numDims; i++) { for(int i = 0; i < _numDims; i++) {
dims[i] = 1; offsets[i] = 1;
for(int j = i+1; j < _numDims; j++) { for(int j = i+1; j < _numDims; j++) {
dims[i] *= lens[j]; offsets[i] *= lens[j];
} }
} }
} }
...@@ -72,44 +72,61 @@ jobject AparapiBuffer::getJavaObject(JNIEnv* env, KernelArg* arg) { ...@@ -72,44 +72,61 @@ jobject AparapiBuffer::getJavaObject(JNIEnv* env, KernelArg* arg) {
} }
AparapiBuffer* AparapiBuffer::flatten(JNIEnv* env, jobject arg, int type) { void AparapiBuffer::flatten(JNIEnv* env, KernelArg* arg) {
int numDims = JNIHelper::getInstanceField<jint>(env, arg, "numDims", IntArg); int numDims = JNIHelper::getInstanceField<jint>(env, arg->javaArg, "numDims", IntArg);
if(numDims == 2 && isBoolean(type)) { if(numDims == 2 && arg->isBoolean()) {
return AparapiBuffer::flattenBoolean2D(env,arg); flattenBoolean2D(env,arg);
} else if(numDims == 2 && isByte(type)) { } else if(numDims == 2 && arg->isByte()) {
return AparapiBuffer::flattenByte2D(env,arg); flattenByte2D(env,arg);
} else if(numDims == 2 && isShort(type)) { } else if(numDims == 2 && arg->isShort()) {
return AparapiBuffer::flattenShort2D(env,arg); flattenShort2D(env,arg);
} else if(numDims == 2 && isInt(type)) { } else if(numDims == 2 && arg->isInt()) {
return AparapiBuffer::flattenInt2D(env,arg); flattenInt2D(env,arg);
} else if(numDims == 2 && isLong(type)) { } else if(numDims == 2 && arg->isLong()) {
return AparapiBuffer::flattenLong2D(env,arg); flattenLong2D(env,arg);
} else if(numDims == 2 && isFloat(type)) { } else if(numDims == 2 && arg->isFloat()) {
return AparapiBuffer::flattenFloat2D(env,arg); flattenFloat2D(env,arg);
} else if(numDims == 2 && isDouble(type)) { } else if(numDims == 2 && arg->isDouble()) {
return AparapiBuffer::flattenDouble2D(env,arg); flattenDouble2D(env,arg);
} else if(numDims == 3 && isBoolean(type)) { } else if(numDims == 3 && arg->isBoolean()) {
return AparapiBuffer::flattenBoolean3D(env,arg); flattenBoolean3D(env,arg);
} else if(numDims == 3 && isByte(type)) { } else if(numDims == 3 && arg->isByte()) {
return AparapiBuffer::flattenByte3D(env,arg); flattenByte3D(env,arg);
} else if(numDims == 3 && isShort(type)) { } else if(numDims == 3 && arg->isShort()) {
return AparapiBuffer::flattenShort3D(env,arg); flattenShort3D(env,arg);
} else if(numDims == 3 && isInt(type)) { } else if(numDims == 3 && arg->isInt()) {
return AparapiBuffer::flattenInt3D(env,arg); flattenInt3D(env,arg);
} else if(numDims == 3 && isLong(type)) { } else if(numDims == 3 && arg->isLong()) {
return AparapiBuffer::flattenLong3D(env,arg); flattenLong3D(env,arg);
} else if(numDims == 3 && isFloat(type)) { } else if(numDims == 3 && arg->isFloat()) {
return AparapiBuffer::flattenFloat3D(env,arg); flattenFloat3D(env,arg);
} else if(numDims == 3 && isDouble(type)) { } else if(numDims == 3 && arg->isDouble()) {
return AparapiBuffer::flattenDouble3D(env,arg); flattenDouble3D(env,arg);
} } else {
return new AparapiBuffer(); fprintf(stderr,"flatten() not understand argument type\n");
}
}
void AparapiBuffer::buildBuffer(void* _data, cl_uint* _dims, cl_uint _numDims, long _lengthInBytes, jobject _javaObject) {
data = _data;
lens = _dims;
numDims = _numDims;
lengthInBytes = _lengthInBytes;
javaObject = _javaObject;
offsets = new cl_uint[_numDims];
for(int i = 0; i < _numDims; i++) {
offsets[i] = 1;
for(int j = i+1; j < _numDims; j++) {
offsets[i] *= lens[j];
}
}
} }
AparapiBuffer* AparapiBuffer::flattenBoolean2D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenBoolean2D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[2]; cl_uint* dims = new cl_uint[2];
dims[0] = env->GetArrayLength((jobjectArray)javaBuffer); dims[0] = env->GetArrayLength((jobjectArray)javaBuffer);
dims[1] = env->GetArrayLength((jbooleanArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0)); dims[1] = env->GetArrayLength((jbooleanArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0));
...@@ -148,14 +165,16 @@ AparapiBuffer* AparapiBuffer::flattenBoolean2D(JNIEnv* env, jobject arg) { ...@@ -148,14 +165,16 @@ AparapiBuffer* AparapiBuffer::flattenBoolean2D(JNIEnv* env, jobject arg) {
array[i*dims[1] + j] = elems[j]; array[i*dims[1] + j] = elems[j];
} }
env->ReleaseBooleanArrayElements(jArray, elems, 0); env->ReleaseBooleanArrayElements(jArray, elems, 0);
// Does DeleteLocalRef required?
// env->DeleteLocalRef(jArray);
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenByte2D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenByte2D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[2]; cl_uint* dims = new cl_uint[2];
dims[0] = env->GetArrayLength((jobjectArray)javaBuffer); dims[0] = env->GetArrayLength((jobjectArray)javaBuffer);
dims[1] = env->GetArrayLength((jbyteArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0)); dims[1] = env->GetArrayLength((jbyteArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0));
...@@ -196,12 +215,12 @@ AparapiBuffer* AparapiBuffer::flattenByte2D(JNIEnv* env, jobject arg) { ...@@ -196,12 +215,12 @@ AparapiBuffer* AparapiBuffer::flattenByte2D(JNIEnv* env, jobject arg) {
env->ReleaseByteArrayElements(jArray, elems, 0); env->ReleaseByteArrayElements(jArray, elems, 0);
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenShort2D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenShort2D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[2]; cl_uint* dims = new cl_uint[2];
dims[0] = env->GetArrayLength((jobjectArray)javaBuffer); dims[0] = env->GetArrayLength((jobjectArray)javaBuffer);
dims[1] = env->GetArrayLength((jshortArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0)); dims[1] = env->GetArrayLength((jshortArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0));
...@@ -242,12 +261,12 @@ AparapiBuffer* AparapiBuffer::flattenShort2D(JNIEnv* env, jobject arg) { ...@@ -242,12 +261,12 @@ AparapiBuffer* AparapiBuffer::flattenShort2D(JNIEnv* env, jobject arg) {
env->ReleaseShortArrayElements(jArray, elems, 0); env->ReleaseShortArrayElements(jArray, elems, 0);
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenInt2D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenInt2D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[2]; cl_uint* dims = new cl_uint[2];
dims[0] = env->GetArrayLength((jobjectArray)javaBuffer); dims[0] = env->GetArrayLength((jobjectArray)javaBuffer);
dims[1] = env->GetArrayLength((jintArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0)); dims[1] = env->GetArrayLength((jintArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0));
...@@ -288,12 +307,12 @@ AparapiBuffer* AparapiBuffer::flattenInt2D(JNIEnv* env, jobject arg) { ...@@ -288,12 +307,12 @@ AparapiBuffer* AparapiBuffer::flattenInt2D(JNIEnv* env, jobject arg) {
env->ReleaseIntArrayElements(jArray, elems, 0); env->ReleaseIntArrayElements(jArray, elems, 0);
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenLong2D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenLong2D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[2]; cl_uint* dims = new cl_uint[2];
dims[0] = env->GetArrayLength((jobjectArray)javaBuffer); dims[0] = env->GetArrayLength((jobjectArray)javaBuffer);
dims[1] = env->GetArrayLength((jlongArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0)); dims[1] = env->GetArrayLength((jlongArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0));
...@@ -334,12 +353,12 @@ AparapiBuffer* AparapiBuffer::flattenLong2D(JNIEnv* env, jobject arg) { ...@@ -334,12 +353,12 @@ AparapiBuffer* AparapiBuffer::flattenLong2D(JNIEnv* env, jobject arg) {
env->ReleaseLongArrayElements(jArray, elems, 0); env->ReleaseLongArrayElements(jArray, elems, 0);
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenFloat2D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenFloat2D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[2]; cl_uint* dims = new cl_uint[2];
dims[0] = env->GetArrayLength((jobjectArray)javaBuffer); dims[0] = env->GetArrayLength((jobjectArray)javaBuffer);
dims[1] = env->GetArrayLength((jfloatArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0)); dims[1] = env->GetArrayLength((jfloatArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0));
...@@ -380,12 +399,12 @@ AparapiBuffer* AparapiBuffer::flattenFloat2D(JNIEnv* env, jobject arg) { ...@@ -380,12 +399,12 @@ AparapiBuffer* AparapiBuffer::flattenFloat2D(JNIEnv* env, jobject arg) {
env->ReleaseFloatArrayElements(jArray, elems, 0); env->ReleaseFloatArrayElements(jArray, elems, 0);
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenDouble2D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenDouble2D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[2]; cl_uint* dims = new cl_uint[2];
dims[0] = env->GetArrayLength((jobjectArray)javaBuffer); dims[0] = env->GetArrayLength((jobjectArray)javaBuffer);
dims[1] = env->GetArrayLength((jdoubleArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0)); dims[1] = env->GetArrayLength((jdoubleArray)env->GetObjectArrayElement((jobjectArray)javaBuffer, 0));
...@@ -426,13 +445,13 @@ AparapiBuffer* AparapiBuffer::flattenDouble2D(JNIEnv* env, jobject arg) { ...@@ -426,13 +445,13 @@ AparapiBuffer* AparapiBuffer::flattenDouble2D(JNIEnv* env, jobject arg) {
env->ReleaseDoubleArrayElements(jArray, elems, 0); env->ReleaseDoubleArrayElements(jArray, elems, 0);
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 2, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenBoolean3D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenBoolean3D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[3]; cl_uint* dims = new cl_uint[3];
jobjectArray j0 = (jobjectArray)javaBuffer; jobjectArray j0 = (jobjectArray)javaBuffer;
jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0); jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0);
...@@ -491,12 +510,12 @@ AparapiBuffer* AparapiBuffer::flattenBoolean3D(JNIEnv* env, jobject arg) { ...@@ -491,12 +510,12 @@ AparapiBuffer* AparapiBuffer::flattenBoolean3D(JNIEnv* env, jobject arg) {
} }
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenByte3D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenByte3D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[3]; cl_uint* dims = new cl_uint[3];
jobjectArray j0 = (jobjectArray)javaBuffer; jobjectArray j0 = (jobjectArray)javaBuffer;
jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0); jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0);
...@@ -555,12 +574,12 @@ AparapiBuffer* AparapiBuffer::flattenByte3D(JNIEnv* env, jobject arg) { ...@@ -555,12 +574,12 @@ AparapiBuffer* AparapiBuffer::flattenByte3D(JNIEnv* env, jobject arg) {
} }
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenShort3D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenShort3D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[3]; cl_uint* dims = new cl_uint[3];
jobjectArray j0 = (jobjectArray)javaBuffer; jobjectArray j0 = (jobjectArray)javaBuffer;
jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0); jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0);
...@@ -619,12 +638,12 @@ AparapiBuffer* AparapiBuffer::flattenShort3D(JNIEnv* env, jobject arg) { ...@@ -619,12 +638,12 @@ AparapiBuffer* AparapiBuffer::flattenShort3D(JNIEnv* env, jobject arg) {
} }
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenInt3D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenInt3D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[3]; cl_uint* dims = new cl_uint[3];
jobjectArray j0 = (jobjectArray)javaBuffer; jobjectArray j0 = (jobjectArray)javaBuffer;
jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0); jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0);
...@@ -683,12 +702,12 @@ AparapiBuffer* AparapiBuffer::flattenInt3D(JNIEnv* env, jobject arg) { ...@@ -683,12 +702,12 @@ AparapiBuffer* AparapiBuffer::flattenInt3D(JNIEnv* env, jobject arg) {
} }
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenLong3D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenLong3D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[3]; cl_uint* dims = new cl_uint[3];
jobjectArray j0 = (jobjectArray)javaBuffer; jobjectArray j0 = (jobjectArray)javaBuffer;
jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0); jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0);
...@@ -747,12 +766,12 @@ AparapiBuffer* AparapiBuffer::flattenLong3D(JNIEnv* env, jobject arg) { ...@@ -747,12 +766,12 @@ AparapiBuffer* AparapiBuffer::flattenLong3D(JNIEnv* env, jobject arg) {
} }
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenFloat3D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenFloat3D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[3]; cl_uint* dims = new cl_uint[3];
jobjectArray j0 = (jobjectArray)javaBuffer; jobjectArray j0 = (jobjectArray)javaBuffer;
jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0); jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0);
...@@ -811,12 +830,12 @@ AparapiBuffer* AparapiBuffer::flattenFloat3D(JNIEnv* env, jobject arg) { ...@@ -811,12 +830,12 @@ AparapiBuffer* AparapiBuffer::flattenFloat3D(JNIEnv* env, jobject arg) {
} }
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer);
} }
AparapiBuffer* AparapiBuffer::flattenDouble3D(JNIEnv* env, jobject arg) { void AparapiBuffer::flattenDouble3D(JNIEnv* env, KernelArg* arg) {
jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg, "javaBuffer", ObjectClassArg); jobject javaBuffer = JNIHelper::getInstanceField<jobject>(env, arg->javaArg, "javaBuffer", ObjectClassArg);
cl_uint* dims = new cl_uint[3]; cl_uint* dims = new cl_uint[3];
jobjectArray j0 = (jobjectArray)javaBuffer; jobjectArray j0 = (jobjectArray)javaBuffer;
jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0); jobjectArray j1 = (jobjectArray)env->GetObjectArrayElement(j0, 0);
...@@ -875,7 +894,7 @@ AparapiBuffer* AparapiBuffer::flattenDouble3D(JNIEnv* env, jobject arg) { ...@@ -875,7 +894,7 @@ AparapiBuffer* AparapiBuffer::flattenDouble3D(JNIEnv* env, jobject arg) {
} }
} }
return new AparapiBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer); buildBuffer((void*)array, (cl_uint*)dims, 3, bitSize, javaBuffer);
} }
...@@ -934,7 +953,7 @@ void AparapiBuffer::inflateBoolean2D(JNIEnv *env, KernelArg* arg) { ...@@ -934,7 +953,7 @@ void AparapiBuffer::inflateBoolean2D(JNIEnv *env, KernelArg* arg) {
#pragma omp parallel for #pragma omp parallel for
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[i][j] = array[i*dims[0] + j]; body[i][j] = array[i*offsets[0] + j];
} }
} }
...@@ -949,7 +968,7 @@ void AparapiBuffer::inflateBoolean2D(JNIEnv *env, KernelArg* arg) { ...@@ -949,7 +968,7 @@ void AparapiBuffer::inflateBoolean2D(JNIEnv *env, KernelArg* arg) {
jbooleanArray jArray = (jbooleanArray)env->GetObjectArrayElement(buffer, i); jbooleanArray jArray = (jbooleanArray)env->GetObjectArrayElement(buffer, i);
jboolean* body = env->GetBooleanArrayElements(jArray,0); jboolean* body = env->GetBooleanArrayElements(jArray,0);
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[j] = array[i*dims[0] + j]; body[j] = array[i*offsets[0] + j];
} }
env->ReleaseBooleanArrayElements(jArray, body, 0); env->ReleaseBooleanArrayElements(jArray, body, 0);
} }
...@@ -971,7 +990,7 @@ void AparapiBuffer::inflateByte2D(JNIEnv *env, KernelArg* arg) { ...@@ -971,7 +990,7 @@ void AparapiBuffer::inflateByte2D(JNIEnv *env, KernelArg* arg) {
#pragma omp parallel for #pragma omp parallel for
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[i][j] = array[i*dims[0] + j]; body[i][j] = array[i*offsets[0] + j];
} }
} }
...@@ -986,7 +1005,7 @@ void AparapiBuffer::inflateByte2D(JNIEnv *env, KernelArg* arg) { ...@@ -986,7 +1005,7 @@ void AparapiBuffer::inflateByte2D(JNIEnv *env, KernelArg* arg) {
jbyteArray jArray = (jbyteArray)env->GetObjectArrayElement(buffer, i); jbyteArray jArray = (jbyteArray)env->GetObjectArrayElement(buffer, i);
jbyte* body = env->GetByteArrayElements(jArray,0); jbyte* body = env->GetByteArrayElements(jArray,0);
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[j] = array[i*dims[0] + j]; body[j] = array[i*offsets[0] + j];
} }
env->ReleaseByteArrayElements(jArray, body, 0); env->ReleaseByteArrayElements(jArray, body, 0);
} }
...@@ -1008,7 +1027,7 @@ void AparapiBuffer::inflateShort2D(JNIEnv *env, KernelArg* arg) { ...@@ -1008,7 +1027,7 @@ void AparapiBuffer::inflateShort2D(JNIEnv *env, KernelArg* arg) {
#pragma omp parallel for #pragma omp parallel for
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[i][j] = array[i*dims[0] + j]; body[i][j] = array[i*offsets[0] + j];
} }
} }
...@@ -1023,7 +1042,7 @@ void AparapiBuffer::inflateShort2D(JNIEnv *env, KernelArg* arg) { ...@@ -1023,7 +1042,7 @@ void AparapiBuffer::inflateShort2D(JNIEnv *env, KernelArg* arg) {
jshortArray jArray = (jshortArray)env->GetObjectArrayElement(buffer, i); jshortArray jArray = (jshortArray)env->GetObjectArrayElement(buffer, i);
jshort* body = env->GetShortArrayElements(jArray,0); jshort* body = env->GetShortArrayElements(jArray,0);
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[j] = array[i*dims[0] + j]; body[j] = array[i*offsets[0] + j];
} }
env->ReleaseShortArrayElements(jArray, body, 0); env->ReleaseShortArrayElements(jArray, body, 0);
} }
...@@ -1045,7 +1064,7 @@ void AparapiBuffer::inflateInt2D(JNIEnv *env, KernelArg* arg) { ...@@ -1045,7 +1064,7 @@ void AparapiBuffer::inflateInt2D(JNIEnv *env, KernelArg* arg) {
#pragma omp parallel for #pragma omp parallel for
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[i][j] = array[i*dims[0] + j]; body[i][j] = array[i*offsets[0] + j];
} }
} }
...@@ -1060,7 +1079,7 @@ void AparapiBuffer::inflateInt2D(JNIEnv *env, KernelArg* arg) { ...@@ -1060,7 +1079,7 @@ void AparapiBuffer::inflateInt2D(JNIEnv *env, KernelArg* arg) {
jintArray jArray = (jintArray)env->GetObjectArrayElement(buffer, i); jintArray jArray = (jintArray)env->GetObjectArrayElement(buffer, i);
jint* body = env->GetIntArrayElements(jArray,0); jint* body = env->GetIntArrayElements(jArray,0);
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[j] = array[i*dims[0] + j]; body[j] = array[i*offsets[0] + j];
} }
env->ReleaseIntArrayElements(jArray, body, 0); env->ReleaseIntArrayElements(jArray, body, 0);
} }
...@@ -1082,7 +1101,7 @@ void AparapiBuffer::inflateLong2D(JNIEnv *env, KernelArg* arg) { ...@@ -1082,7 +1101,7 @@ void AparapiBuffer::inflateLong2D(JNIEnv *env, KernelArg* arg) {
#pragma omp parallel for #pragma omp parallel for
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[i][j] = array[i*dims[0] + j]; body[i][j] = array[i*offsets[0] + j];
} }
} }
...@@ -1097,7 +1116,7 @@ void AparapiBuffer::inflateLong2D(JNIEnv *env, KernelArg* arg) { ...@@ -1097,7 +1116,7 @@ void AparapiBuffer::inflateLong2D(JNIEnv *env, KernelArg* arg) {
jlongArray jArray = (jlongArray)env->GetObjectArrayElement(buffer, i); jlongArray jArray = (jlongArray)env->GetObjectArrayElement(buffer, i);
jlong* body = env->GetLongArrayElements(jArray,0); jlong* body = env->GetLongArrayElements(jArray,0);
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[j] = array[i*dims[0] + j]; body[j] = array[i*offsets[0] + j];
} }
env->ReleaseLongArrayElements(jArray, body, 0); env->ReleaseLongArrayElements(jArray, body, 0);
} }
...@@ -1119,7 +1138,7 @@ void AparapiBuffer::inflateFloat2D(JNIEnv *env, KernelArg* arg) { ...@@ -1119,7 +1138,7 @@ void AparapiBuffer::inflateFloat2D(JNIEnv *env, KernelArg* arg) {
#pragma omp parallel for #pragma omp parallel for
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[i][j] = array[i*dims[0] + j]; body[i][j] = array[i*offsets[0] + j];
} }
} }
...@@ -1134,7 +1153,7 @@ void AparapiBuffer::inflateFloat2D(JNIEnv *env, KernelArg* arg) { ...@@ -1134,7 +1153,7 @@ void AparapiBuffer::inflateFloat2D(JNIEnv *env, KernelArg* arg) {
jfloatArray jArray = (jfloatArray)env->GetObjectArrayElement(buffer, i); jfloatArray jArray = (jfloatArray)env->GetObjectArrayElement(buffer, i);
jfloat* body = env->GetFloatArrayElements(jArray,0); jfloat* body = env->GetFloatArrayElements(jArray,0);
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[j] = array[i*dims[0] + j]; body[j] = array[i*offsets[0] + j];
} }
env->ReleaseFloatArrayElements(jArray, body, 0); env->ReleaseFloatArrayElements(jArray, body, 0);
} }
...@@ -1156,7 +1175,7 @@ void AparapiBuffer::inflateDouble2D(JNIEnv *env, KernelArg* arg) { ...@@ -1156,7 +1175,7 @@ void AparapiBuffer::inflateDouble2D(JNIEnv *env, KernelArg* arg) {
#pragma omp parallel for #pragma omp parallel for
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[i][j] = array[i*dims[0] + j]; body[i][j] = array[i*offsets[0] + j];
} }
} }
...@@ -1171,7 +1190,7 @@ void AparapiBuffer::inflateDouble2D(JNIEnv *env, KernelArg* arg) { ...@@ -1171,7 +1190,7 @@ void AparapiBuffer::inflateDouble2D(JNIEnv *env, KernelArg* arg) {
jdoubleArray jArray = (jdoubleArray)env->GetObjectArrayElement(buffer, i); jdoubleArray jArray = (jdoubleArray)env->GetObjectArrayElement(buffer, i);
jdouble* body = env->GetDoubleArrayElements(jArray,0); jdouble* body = env->GetDoubleArrayElements(jArray,0);
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
body[j] = array[i*dims[0] + j]; body[j] = array[i*offsets[0] + j];
} }
env->ReleaseDoubleArrayElements(jArray, body, 0); env->ReleaseDoubleArrayElements(jArray, body, 0);
} }
...@@ -1200,7 +1219,7 @@ void AparapiBuffer::inflateBoolean3D(JNIEnv *env, KernelArg* arg) { ...@@ -1200,7 +1219,7 @@ void AparapiBuffer::inflateBoolean3D(JNIEnv *env, KernelArg* arg) {
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[i][j][k] = array[i*dims[0] + j*dims[1] + k]; body[i][j][k] = array[i*offsets[0] + j*offsets[1] + k];
} }
} }
} }
...@@ -1222,7 +1241,7 @@ void AparapiBuffer::inflateBoolean3D(JNIEnv *env, KernelArg* arg) { ...@@ -1222,7 +1241,7 @@ void AparapiBuffer::inflateBoolean3D(JNIEnv *env, KernelArg* arg) {
jbooleanArray jArray = (jbooleanArray)env->GetObjectArrayElement(jrow, j); jbooleanArray jArray = (jbooleanArray)env->GetObjectArrayElement(jrow, j);
jboolean* body = env->GetBooleanArrayElements(jArray,0); jboolean* body = env->GetBooleanArrayElements(jArray,0);
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[k] = array[i*dims[0] + j*dims[1] + k]; body[k] = array[i*offsets[0] + j*offsets[1] + k];
} }
env->ReleaseBooleanArrayElements(jArray, body, 0); env->ReleaseBooleanArrayElements(jArray, body, 0);
} }
...@@ -1251,7 +1270,7 @@ void AparapiBuffer::inflateByte3D(JNIEnv *env, KernelArg* arg) { ...@@ -1251,7 +1270,7 @@ void AparapiBuffer::inflateByte3D(JNIEnv *env, KernelArg* arg) {
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[i][j][k] = array[i*dims[0] + j*dims[1] + k]; body[i][j][k] = array[i*offsets[0] + j*offsets[1] + k];
} }
} }
} }
...@@ -1273,7 +1292,7 @@ void AparapiBuffer::inflateByte3D(JNIEnv *env, KernelArg* arg) { ...@@ -1273,7 +1292,7 @@ void AparapiBuffer::inflateByte3D(JNIEnv *env, KernelArg* arg) {
jbyteArray jArray = (jbyteArray)env->GetObjectArrayElement(jrow, j); jbyteArray jArray = (jbyteArray)env->GetObjectArrayElement(jrow, j);
jbyte* body = env->GetByteArrayElements(jArray,0); jbyte* body = env->GetByteArrayElements(jArray,0);
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[k] = array[i*dims[0] + j*dims[1] + k]; body[k] = array[i*offsets[0] + j*offsets[1] + k];
} }
env->ReleaseByteArrayElements(jArray, body, 0); env->ReleaseByteArrayElements(jArray, body, 0);
} }
...@@ -1302,7 +1321,7 @@ void AparapiBuffer::inflateShort3D(JNIEnv *env, KernelArg* arg) { ...@@ -1302,7 +1321,7 @@ void AparapiBuffer::inflateShort3D(JNIEnv *env, KernelArg* arg) {
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[i][j][k] = array[i*dims[0] + j*dims[1] + k]; body[i][j][k] = array[i*offsets[0] + j*offsets[1] + k];
} }
} }
} }
...@@ -1324,7 +1343,7 @@ void AparapiBuffer::inflateShort3D(JNIEnv *env, KernelArg* arg) { ...@@ -1324,7 +1343,7 @@ void AparapiBuffer::inflateShort3D(JNIEnv *env, KernelArg* arg) {
jshortArray jArray = (jshortArray)env->GetObjectArrayElement(jrow, j); jshortArray jArray = (jshortArray)env->GetObjectArrayElement(jrow, j);
jshort* body = env->GetShortArrayElements(jArray,0); jshort* body = env->GetShortArrayElements(jArray,0);
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[k] = array[i*dims[0] + j*dims[1] + k]; body[k] = array[i*offsets[0] + j*offsets[1] + k];
} }
env->ReleaseShortArrayElements(jArray, body, 0); env->ReleaseShortArrayElements(jArray, body, 0);
} }
...@@ -1353,7 +1372,7 @@ void AparapiBuffer::inflateInt3D(JNIEnv *env, KernelArg* arg) { ...@@ -1353,7 +1372,7 @@ void AparapiBuffer::inflateInt3D(JNIEnv *env, KernelArg* arg) {
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[i][j][k] = array[i*dims[0] + j*dims[1] + k]; body[i][j][k] = array[i*offsets[0] + j*offsets[1] + k];
} }
} }
} }
...@@ -1375,7 +1394,7 @@ void AparapiBuffer::inflateInt3D(JNIEnv *env, KernelArg* arg) { ...@@ -1375,7 +1394,7 @@ void AparapiBuffer::inflateInt3D(JNIEnv *env, KernelArg* arg) {
jintArray jArray = (jintArray)env->GetObjectArrayElement(jrow, j); jintArray jArray = (jintArray)env->GetObjectArrayElement(jrow, j);
jint* body = env->GetIntArrayElements(jArray,0); jint* body = env->GetIntArrayElements(jArray,0);
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[k] = array[i*dims[0] + j*dims[1] + k]; body[k] = array[i*offsets[0] + j*offsets[1] + k];
} }
env->ReleaseIntArrayElements(jArray, body, 0); env->ReleaseIntArrayElements(jArray, body, 0);
} }
...@@ -1404,7 +1423,7 @@ void AparapiBuffer::inflateLong3D(JNIEnv *env, KernelArg* arg) { ...@@ -1404,7 +1423,7 @@ void AparapiBuffer::inflateLong3D(JNIEnv *env, KernelArg* arg) {
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[i][j][k] = array[i*dims[0] + j*dims[1] + k]; body[i][j][k] = array[i*offsets[0] + j*offsets[1] + k];
} }
} }
} }
...@@ -1426,7 +1445,7 @@ void AparapiBuffer::inflateLong3D(JNIEnv *env, KernelArg* arg) { ...@@ -1426,7 +1445,7 @@ void AparapiBuffer::inflateLong3D(JNIEnv *env, KernelArg* arg) {
jlongArray jArray = (jlongArray)env->GetObjectArrayElement(jrow, j); jlongArray jArray = (jlongArray)env->GetObjectArrayElement(jrow, j);
jlong* body = env->GetLongArrayElements(jArray,0); jlong* body = env->GetLongArrayElements(jArray,0);
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[k] = array[i*dims[0] + j*dims[1] + k]; body[k] = array[i*offsets[0] + j*offsets[1] + k];
} }
env->ReleaseLongArrayElements(jArray, body, 0); env->ReleaseLongArrayElements(jArray, body, 0);
} }
...@@ -1455,7 +1474,7 @@ void AparapiBuffer::inflateFloat3D(JNIEnv *env, KernelArg* arg) { ...@@ -1455,7 +1474,7 @@ void AparapiBuffer::inflateFloat3D(JNIEnv *env, KernelArg* arg) {
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[i][j][k] = array[i*dims[0] + j*dims[1] + k]; body[i][j][k] = array[i*offsets[0] + j*offsets[1] + k];
} }
} }
} }
...@@ -1477,7 +1496,7 @@ void AparapiBuffer::inflateFloat3D(JNIEnv *env, KernelArg* arg) { ...@@ -1477,7 +1496,7 @@ void AparapiBuffer::inflateFloat3D(JNIEnv *env, KernelArg* arg) {
jfloatArray jArray = (jfloatArray)env->GetObjectArrayElement(jrow, j); jfloatArray jArray = (jfloatArray)env->GetObjectArrayElement(jrow, j);
jfloat* body = env->GetFloatArrayElements(jArray,0); jfloat* body = env->GetFloatArrayElements(jArray,0);
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[k] = array[i*dims[0] + j*dims[1] + k]; body[k] = array[i*offsets[0] + j*offsets[1] + k];
} }
env->ReleaseFloatArrayElements(jArray, body, 0); env->ReleaseFloatArrayElements(jArray, body, 0);
} }
...@@ -1506,7 +1525,7 @@ void AparapiBuffer::inflateDouble3D(JNIEnv *env, KernelArg* arg) { ...@@ -1506,7 +1525,7 @@ void AparapiBuffer::inflateDouble3D(JNIEnv *env, KernelArg* arg) {
for(int i = 0; i < lens[0]; i++) { for(int i = 0; i < lens[0]; i++) {
for(int j = 0; j < lens[1]; j++) { for(int j = 0; j < lens[1]; j++) {
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[i][j][k] = array[i*dims[0] + j*dims[1] + k]; body[i][j][k] = array[i*offsets[0] + j*offsets[1] + k];
} }
} }
} }
...@@ -1528,7 +1547,7 @@ void AparapiBuffer::inflateDouble3D(JNIEnv *env, KernelArg* arg) { ...@@ -1528,7 +1547,7 @@ void AparapiBuffer::inflateDouble3D(JNIEnv *env, KernelArg* arg) {
jdoubleArray jArray = (jdoubleArray)env->GetObjectArrayElement(jrow, j); jdoubleArray jArray = (jdoubleArray)env->GetObjectArrayElement(jrow, j);
jdouble* body = env->GetDoubleArrayElements(jArray,0); jdouble* body = env->GetDoubleArrayElements(jArray,0);
for(int k = 0; k < lens[2]; k++) { for(int k = 0; k < lens[2]; k++) {
body[k] = array[i*dims[0] + j*dims[1] + k]; body[k] = array[i*offsets[0] + j*offsets[1] + k];
} }
env->ReleaseDoubleArrayElements(jArray, body, 0); env->ReleaseDoubleArrayElements(jArray, body, 0);
} }
...@@ -1537,7 +1556,7 @@ void AparapiBuffer::inflateDouble3D(JNIEnv *env, KernelArg* arg) { ...@@ -1537,7 +1556,7 @@ void AparapiBuffer::inflateDouble3D(JNIEnv *env, KernelArg* arg) {
void AparapiBuffer::deleteBuffer(KernelArg* arg) void AparapiBuffer::deleteBuffer(KernelArg* arg)
{ {
delete[] dims; delete[] offsets;
delete[] lens; delete[] lens;
if(arg->isBoolean()) { if(arg->isBoolean()) {
delete[] (jboolean*)data; delete[] (jboolean*)data;
......
com.amd.aparapi.jni/src/cpp/runKernel/AparapiBuffer.h
+ 23
− 21
View file @ 6319738b
...@@ -70,10 +70,12 @@ private: ...@@ -70,10 +70,12 @@ private:
return (type&com_amd_aparapi_internal_jni_KernelRunnerJNI_ARG_SHORT); return (type&com_amd_aparapi_internal_jni_KernelRunnerJNI_ARG_SHORT);
} }
void buildBuffer(void* _data, cl_uint* _dims, cl_uint _numDims, long _lengthInBytes, jobject _javaObject);
public: public:
jobject javaObject; // The java array that this arg is mapped to jobject javaObject; // The java array that this arg is mapped to
cl_uint numDims; // sizes of dimensions of the object (array lengths for ND arrays) cl_uint numDims; // number of dimensions of the object (array lengths for ND arrays)
cl_uint* dims; // sizes of offsets of the object (first element offset in ND arrays) cl_uint* offsets; // offsets of the next element in ND arrays)
cl_uint* lens; // sizes of dimensions of the object (array lengths for ND arrays) cl_uint* lens; // sizes of dimensions of the object (array lengths for ND arrays)
jint lengthInBytes; // bytes in the array or directBuf jint lengthInBytes; // bytes in the array or directBuf
cl_mem mem; // the opencl buffer cl_mem mem; // the opencl buffer
...@@ -87,25 +89,25 @@ public: ...@@ -87,25 +89,25 @@ public:
void deleteBuffer(KernelArg* arg); void deleteBuffer(KernelArg* arg);
static AparapiBuffer* flatten(JNIEnv *env, jobject arg, int type); void flatten(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenBoolean2D(JNIEnv *env, jobject arg); void flattenBoolean2D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenChar2D(JNIEnv *env, jobject arg); void flattenChar2D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenByte2D(JNIEnv *env, jobject arg); void flattenByte2D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenShort2D(JNIEnv *env, jobject arg); void flattenShort2D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenInt2D(JNIEnv *env, jobject arg); void flattenInt2D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenLong2D(JNIEnv *env, jobject arg); void flattenLong2D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenFloat2D(JNIEnv *env, jobject arg); void flattenFloat2D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenDouble2D(JNIEnv *env, jobject arg); void flattenDouble2D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenBoolean3D(JNIEnv *env, jobject arg); void flattenBoolean3D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenChar3D(JNIEnv *env, jobject arg); void flattenChar3D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenByte3D(JNIEnv *env, jobject arg); void flattenByte3D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenShort3D(JNIEnv *env, jobject arg); void flattenShort3D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenInt3D(JNIEnv *env, jobject arg); void flattenInt3D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenLong3D(JNIEnv *env, jobject arg); void flattenLong3D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenFloat3D(JNIEnv *env, jobject arg); void flattenFloat3D(JNIEnv *env, KernelArg* arg);
static AparapiBuffer* flattenDouble3D(JNIEnv *env, jobject arg); void flattenDouble3D(JNIEnv *env, KernelArg* arg);
void inflate(JNIEnv *env, KernelArg* arg); void inflate(JNIEnv *env, KernelArg* arg);
......
com.amd.aparapi.jni/src/cpp/runKernel/JNIContext.cpp
+ 35
− 16
View file @ 6319738b
...@@ -21,8 +21,7 @@ JNIContext::JNIContext(JNIEnv *jenv, jobject _kernelObject, jobject _openCLDevic ...@@ -21,8 +21,7 @@ JNIContext::JNIContext(JNIEnv *jenv, jobject _kernelObject, jobject _openCLDevic
deviceId = OpenCLDevice::getDeviceId(jenv, openCLDeviceObject); deviceId = OpenCLDevice::getDeviceId(jenv, openCLDeviceObject);
cl_device_type returnedDeviceType; cl_device_type returnedDeviceType;
clGetDeviceInfo(deviceId, CL_DEVICE_TYPE, sizeof(returnedDeviceType), &returnedDeviceType, NULL); clGetDeviceInfo(deviceId, CL_DEVICE_TYPE, sizeof(returnedDeviceType), &returnedDeviceType, NULL);
//fprintf(stderr, "device[%d] CL_DEVICE_TYPE = %x\n", deviceId, returnedDeviceType); //fprintf(stderr, "device[%p] CL_DEVICE_TYPE = %x\n", deviceId, returnedDeviceType);
cl_context_properties cps[3] = { CL_CONTEXT_PLATFORM, (cl_context_properties)platformId, 0 }; cl_context_properties cps[3] = { CL_CONTEXT_PLATFORM, (cl_context_properties)platformId, 0 };
cl_context_properties* cprops = (NULL == platformId) ? NULL : cps; cl_context_properties* cprops = (NULL == platformId) ? NULL : cps;
...@@ -34,7 +33,7 @@ JNIContext::JNIContext(JNIEnv *jenv, jobject _kernelObject, jobject _openCLDevic ...@@ -34,7 +33,7 @@ JNIContext::JNIContext(JNIEnv *jenv, jobject _kernelObject, jobject _openCLDevic
} }
void JNIContext::dispose(JNIEnv *jenv, Config* config) { void JNIContext::dispose(JNIEnv *jenv, Config* config) {
//fprintf(stdout, "dispose()\n"); //fprintf(stderr, "dispose()\n");
cl_int status = CL_SUCCESS; cl_int status = CL_SUCCESS;
jenv->DeleteGlobalRef(kernelObject); jenv->DeleteGlobalRef(kernelObject);
jenv->DeleteGlobalRef(kernelClass); jenv->DeleteGlobalRef(kernelClass);
...@@ -69,21 +68,41 @@ void JNIContext::dispose(JNIEnv *jenv, Config* config) { ...@@ -69,21 +68,41 @@ void JNIContext::dispose(JNIEnv *jenv, Config* config) {
for (int i=0; i< argc; i++){ for (int i=0; i< argc; i++){
KernelArg *arg = args[i]; KernelArg *arg = args[i];
if (!arg->isPrimitive()){ if (!arg->isPrimitive()){
if (arg->arrayBuffer != NULL){ if (arg->isArray()) {
if (arg->arrayBuffer->mem != 0){ if (arg->arrayBuffer != NULL){
if (config->isTrackingOpenCLResources()){ if (arg->arrayBuffer->mem != 0){
memList.remove((cl_mem)arg->arrayBuffer->mem, __LINE__, __FILE__); if (config->isTrackingOpenCLResources()){
memList.remove((cl_mem)arg->arrayBuffer->mem, __LINE__, __FILE__);
}
status = clReleaseMemObject((cl_mem)arg->arrayBuffer->mem);
//fprintf(stdout, "dispose arg %d %0lx\n", i, arg->arrayBuffer->mem);
CLException::checkCLError(status, "clReleaseMemObject()");
arg->arrayBuffer->mem = (cl_mem)0;
}
if (arg->arrayBuffer->javaArray != NULL) {
jenv->DeleteWeakGlobalRef((jweak) arg->arrayBuffer->javaArray);
} }
status = clReleaseMemObject((cl_mem)arg->arrayBuffer->mem); delete arg->arrayBuffer;
//fprintf(stdout, "dispose arg %d %0lx\n", i, arg->arrayBuffer->mem); arg->arrayBuffer = NULL;
CLException::checkCLError(status, "clReleaseMemObject()");
arg->arrayBuffer->mem = (cl_mem)0;
} }
if (arg->arrayBuffer->javaArray != NULL) { } else if (arg->isAparapiBuffer()) {
jenv->DeleteWeakGlobalRef((jweak) arg->arrayBuffer->javaArray); if (arg->aparapiBuffer != NULL){
if (arg->aparapiBuffer->mem != 0){
if (config->isTrackingOpenCLResources()){
memList.remove((cl_mem)arg->aparapiBuffer->mem, __LINE__, __FILE__);
}
status = clReleaseMemObject((cl_mem)arg->aparapiBuffer->mem);
//fprintf(stdout, "dispose arg %d %0lx\n", i, arg->aparapiBuffer->mem);
CLException::checkCLError(status, "clReleaseMemObject()");
arg->aparapiBuffer->mem = (cl_mem)0;
}
if (arg->aparapiBuffer->javaObject != NULL) {
jenv->DeleteWeakGlobalRef((jweak) arg->aparapiBuffer->javaObject);
}
delete arg->aparapiBuffer;
arg->aparapiBuffer = NULL;
} }
delete arg->arrayBuffer;
arg->arrayBuffer = NULL;
} }
} }
if (arg->name != NULL){ if (arg->name != NULL){
...@@ -93,7 +112,7 @@ void JNIContext::dispose(JNIEnv *jenv, Config* config) { ...@@ -93,7 +112,7 @@ void JNIContext::dispose(JNIEnv *jenv, Config* config) {
jenv->DeleteGlobalRef((jobject) arg->javaArg); jenv->DeleteGlobalRef((jobject) arg->javaArg);
} }
delete arg; arg=args[i]=NULL; delete arg; arg=args[i]=NULL;
} } // for
delete[] args; args=NULL; delete[] args; args=NULL;
// do we need to call clReleaseEvent on any of these that are still retained.... // do we need to call clReleaseEvent on any of these that are still retained....
......
com.amd.aparapi.jni/src/cpp/runKernel/KernelArg.cpp
+ 1
− 1
View file @ 6319738b
...@@ -35,7 +35,7 @@ KernelArg::KernelArg(JNIEnv *jenv, JNIContext *jniContext, jobject argObj): ...@@ -35,7 +35,7 @@ KernelArg::KernelArg(JNIEnv *jenv, JNIContext *jniContext, jobject argObj):
if (isArray()){ if (isArray()){
arrayBuffer = new ArrayBuffer(); arrayBuffer = new ArrayBuffer();
} else if(isAparapiBuffer()) { } else if(isAparapiBuffer()) {
aparapiBuffer = AparapiBuffer::flatten(jenv, argObj, type); aparapiBuffer = new AparapiBuffer();
} }
} }
......
com.amd.aparapi.jni/src/cpp/runKernel/KernelArg.h
+ 1
− 1
View file @ 6319738b
...@@ -64,7 +64,7 @@ class KernelArg{ ...@@ -64,7 +64,7 @@ class KernelArg{
static jfieldID javaArrayFieldID; static jfieldID javaArrayFieldID;
public: public:
JNIContext *jniContext; JNIContext *jniContext;
jobject argObj; // the Java KernelRunner.KernelArg object that we are mirroring. jobject argObj; // the Java KernelRunner.KernelArg object that we are mirroring. Do not use it outside constructor due to GC. Use javaArg instead.
jobject javaArg; // global reference to the corresponding java KernelArg object we grabbed our own global reference so that the object won't be collected until we dispose! jobject javaArg; // global reference to the corresponding java KernelArg object we grabbed our own global reference so that the object won't be collected until we dispose!
char *name; // used for debugging printfs char *name; // used for debugging printfs
jint type; // a bit mask determining the type of this arg jint type; // a bit mask determining the type of this arg
......
com.amd.aparapi/src/java/com/amd/aparapi/internal/kernel/KernelRunner.java
+ 42
− 35
View file @ 6319738b
...@@ -986,6 +986,34 @@ public class KernelRunner extends KernelRunnerJNI{ ...@@ -986,6 +986,34 @@ public class KernelRunner extends KernelRunnerJNI{
arg.setArray(newArrayRef); arg.setArray(newArrayRef);
assert arg.getArray() != null : "null array ref"; assert arg.getArray() != null : "null array ref";
} else if ((arg.getType() & ARG_APARAPI_BUFFER) != 0) {
// TODO: check if the 2D/3D array is changed.
// can Arrays.equals help?
needsSync = true; // Always need syn
Object buffer = new Object();
try {
buffer = arg.getField().get(kernel);
} catch (IllegalAccessException e) {
e.printStackTrace();
}
int numDims = arg.getNumDims();
Object subBuffer = buffer;
int[] dims = new int[numDims];
for (int d = 0; d < numDims - 1; d++) {
dims[d] = Array.getLength(subBuffer);
subBuffer = Array.get(subBuffer, 0);
}
dims[numDims - 1] = Array.getLength(subBuffer);
arg.setDims(dims);
int primitiveSize = getPrimitiveSize(arg.getType());
int totalElements = 1;
for (int d = 0; d < numDims; d++) {
totalElements *= dims[d];
}
arg.setJavaBuffer(buffer);
arg.setSizeInBytes(totalElements * primitiveSize);
arg.setArray(buffer);
} }
} catch (final IllegalArgumentException e) { } catch (final IllegalArgumentException e) {
e.printStackTrace(); e.printStackTrace();
...@@ -1619,53 +1647,32 @@ public class KernelRunner extends KernelRunnerJNI{ ...@@ -1619,53 +1647,32 @@ public class KernelRunner extends KernelRunnerJNI{
while (type.getName().charAt(numDims) == '[') { while (type.getName().charAt(numDims) == '[') {
numDims++; numDims++;
} }
Object buffer = new Object();
try {
buffer = arg.getField().get(kernel);
} catch (IllegalAccessException e) {
e.printStackTrace();
}
arg.setJavaBuffer(buffer);
arg.setNumDims(numDims); arg.setNumDims(numDims);
Object subBuffer = buffer; arg.setJavaBuffer(null); // will get updated in updateKernelArrayRefs
int[] dims = new int[numDims]; arg.setArray(null); // will get updated in updateKernelArrayRefs
for (int i = 0; i < numDims - 1; i++) {
dims[i] = Array.getLength(subBuffer); Class<?> elementType = arg.getField().getType();
subBuffer = Array.get(subBuffer, 0); while (elementType.isArray()) {
elementType = elementType.getComponentType();
} }
dims[numDims - 1] = Array.getLength(subBuffer);
arg.setDims(dims);
if (subBuffer.getClass().isAssignableFrom(float[].class)) { if (elementType.isAssignableFrom(float.class)) {
arg.setType(arg.getType() | ARG_FLOAT); arg.setType(arg.getType() | ARG_FLOAT);
} } else if (elementType.isAssignableFrom(int.class)) {
if (subBuffer.getClass().isAssignableFrom(int[].class)) {
arg.setType(arg.getType() | ARG_INT); arg.setType(arg.getType() | ARG_INT);
} } else if (elementType.isAssignableFrom(boolean.class)) {
if (subBuffer.getClass().isAssignableFrom(boolean[].class)) {
arg.setType(arg.getType() | ARG_BOOLEAN); arg.setType(arg.getType() | ARG_BOOLEAN);
} } else if (elementType.isAssignableFrom(byte.class)) {
if (subBuffer.getClass().isAssignableFrom(byte[].class)) {
arg.setType(arg.getType() | ARG_BYTE); arg.setType(arg.getType() | ARG_BYTE);
} } else if (elementType.isAssignableFrom(char.class)) {
if (subBuffer.getClass().isAssignableFrom(char[].class)) {
arg.setType(arg.getType() | ARG_CHAR); arg.setType(arg.getType() | ARG_CHAR);
} } else if (elementType.isAssignableFrom(double.class)) {
if (subBuffer.getClass().isAssignableFrom(double[].class)) {
arg.setType(arg.getType() | ARG_DOUBLE); arg.setType(arg.getType() | ARG_DOUBLE);
} } else if (elementType.isAssignableFrom(long.class)) {
if (subBuffer.getClass().isAssignableFrom(long[].class)) {
arg.setType(arg.getType() | ARG_LONG); arg.setType(arg.getType() | ARG_LONG);
} } else if (elementType.isAssignableFrom(short.class)) {
if (subBuffer.getClass().isAssignableFrom(short[].class)) {
arg.setType(arg.getType() | ARG_SHORT); arg.setType(arg.getType() | ARG_SHORT);
} }
int primitiveSize = getPrimitiveSize(arg.getType());
int totalElements = 1;
for (int i = 0; i < numDims; i++) {
totalElements *= dims[i];
}
arg.setSizeInBytes(totalElements * primitiveSize);
} }
private final Set<Object> puts = new HashSet<Object>(); private final Set<Object> puts = new HashSet<Object>();
......
0% or .
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment