A demo script for running opencl on android, it should ouput
1 0 or 1 delta.
Note that google does not want to you to run opencl on android
so that this script may not work even if your android device
is the newest flagship model. Search for "opencl" in google play
for some apps that can check for opencl capability.
Please report if your devices support opencl but this script failed.
--
regards,
====================================================
GPG key 1024D/4434BAB3 2008-08-24
gpg --keyserver subkeys.pgp.net --recv-keys 4434BAB3
gpg --keyserver subkeys.pgp.net --armor --export 4434BAB3
cocurrent 'test'
SZI=: IF64{4 8
CL_MEM_WRITE_ONLY=: 2
CL_MEM_READ_ONLY=: 4
CL_TRUE=: 1
i2d=: 3 : 0
_2 fc 2 ic ,y,.0
)
3 : 0''
libopencl=: ''
for_f. <;._1' libOpenCL.so libPVROCL.so libllvm-a3xx.so libGLES_mali.so' do.
if. fexist f1=. <'/system/lib/', >f do.
libopencl=: f1 break.
elseif. fexist f1=. <'/system/vendor/lib/', >f do.
libopencl=: f1 break.
elseif. fexist f1=. <'/system/vendor/lib/egl/', >f do.
libopencl=: f1 break.
end.
end.
libopencl=: >libopencl
assert. *#libopencl [ 'opencl not supported'
''
)
clBuildProgram=: (libopencl,' clBuildProgram > i x i *x *c x x')&cd
clCreateBuffer=: (libopencl,' clCreateBuffer > x x d x x *i')&cd
clCreateCommandQueue=: (libopencl,' clCreateCommandQueue > x x x d *i')&cd
clCreateContext=: (libopencl,' clCreateContext > x *x i *x x x *i')&cd
clCreateKernel=: (libopencl,' clCreateKernel > x x *c *i')&cd
clCreateProgramWithSource=: (libopencl,' clCreateProgramWithSource > x x i *x
*x *i')&cd
clEnqueueNDRangeKernel=: (libopencl,' clEnqueueNDRangeKernel > i x x i *x *x *x
i *x *x')&cd
clEnqueueReadBuffer_f=: (libopencl,' clEnqueueReadBuffer > i x x i x x *f i *x
*x')&cd
clEnqueueWriteBuffer_f=: (libopencl,' clEnqueueWriteBuffer > i x x i x x *f i
*x *x')&cd
clGetDeviceIDs=: (libopencl,' clGetDeviceIDs > i x d i *x *i')&cd
clGetPlatformIDs=: (libopencl,' clGetPlatformIDs > i i *x *i')&cd
clReleaseCommandQueue=: (libopencl,' clReleaseCommandQueue > i x')&cd
clReleaseContext=: (libopencl,' clReleaseContext > i x')&cd
clReleaseKernel=: (libopencl,' clReleaseKernel > i x')&cd
clReleaseMemObject=: (libopencl,' clReleaseMemObject > i x')&cd
clReleaseProgram=: (libopencl,' clReleaseProgram > i x')&cd
clSetKernelArg=: (libopencl,' clSetKernelArg > i x i x x')&cd
kernel=: 0 : 0
__kernel void vecAdd( __global float *a,
__global float *b,
__global float *c,
const unsigned int n)
{
//Get our global thread ID
int id = get_global_id(0);
//Make sure we do not go out of bounds
if (id < n)
c[id] = a[id] + b[id];
}
)
test=: 3 : 0
clGetPlatformIDs 1; (platform=. ,_1); <<0
clGetDeviceIDs ({.platform); ({.@i2d _1); 1; (dev=. ,_1); <<0
context=. clCreateContext (<0); 1; dev; 0; 0; <<0
program=. clCreateProgramWithSource context; 1; (,15!:14 <'kernel');(,#kernel);
<<0
clBuildProgram program; 1; dev; ''; 0; 0
queue=. clCreateCommandQueue context; ({.dev); ({.@i2d 0); <<0
kernel=. clCreateKernel program; 'vecAdd'; <<0
n=. 100000
bytes=. n * 4 NB. float
h_a=. *: 1&o. i.n
h_b=. *: 2&o. i.n
h_c=. n$(1.5-1.5)
d_a=. clCreateBuffer context; ({.@i2d CL_MEM_READ_ONLY); bytes; 0; <<0
d_b=. clCreateBuffer context; ({.@i2d CL_MEM_READ_ONLY); bytes; 0; <<0
d_c=. clCreateBuffer context; ({.@i2d CL_MEM_WRITE_ONLY); bytes; 0; <<0
localSize=. , 1
globalSize=. , localSize*(>.n%localSize)
clEnqueueWriteBuffer_f queue; d_a; CL_TRUE; 0; bytes; h_a; 0; (<0) ; <<0
clEnqueueWriteBuffer_f queue; d_b; CL_TRUE; 0; bytes; h_b; 0; (<0) ; <<0
clSetKernelArg kernel; 0; SZI; (15!:14<'d_a')
clSetKernelArg kernel; 1; SZI; (15!:14<'d_b')
clSetKernelArg kernel; 2; SZI; (15!:14<'d_c')
clSetKernelArg kernel; 3; 4; (15!:14<'n')
clEnqueueNDRangeKernel queue; kernel; 1; (<0); globalSize; localSize; 0; (<0);
<<0
clEnqueueReadBuffer_f queue; d_c; CL_TRUE; 0; bytes; h_c; 0; (<0) ; <<0
smoutput 1 (],-) n%~+/h_c
clReleaseMemObject d_a
clReleaseMemObject d_b
clReleaseMemObject d_c
clReleaseKernel kernel
clReleaseCommandQueue queue
clReleaseProgram program
clReleaseContext context
smoutput 'finish'
''
)
test''
----------------------------------------------------------------------
For information about J forums see http://www.jsoftware.com/forums.htm
