高通平台下安卓opencl小例子高通平台下安卓opencl小例子

http://blog.csdn.net/wcj0626/article/details/26272019

先到高通的qdn下载adreno GPU SDK，里面有OpenCL的例子。
https://developer.qualcomm.com/software/adreno-gpu-sdk

例子在以下目录：
AdrenoSDK\Development\Samples\OpenCL

1、从高通官网下载Adreon SDK（需要注册为会员）
2、解压以后，把文件夹Development/Inc/内的CL文件夹拷贝到：Android-ndk-r9d/platforms/android-19/arch-arm/usr/include/（请自行选择拷贝到哪个版本下）
3、连接手机到电脑，在终端运行：adb pull /system/vendor/lib/libOpenCL.so,此时会把手机里的libOpenCL.so拷贝到电脑当前文件夹下，把该文件拷贝到：android-ndk-r9d/platforms/android-19/arch-arm/usr/lib/（请自行选择拷贝到哪个版本下）
4、新建安卓工程，NDK开发。不知道NDK怎么弄，请移步这里
5、Java代码：
【高通平台下安卓opencl小例子】
[java]view plain copy

public class MainActivity extends ActionBarActivity {
static
{
System.loadLibrary("ocl");
}
public native Stringtestopencl();
public native String getPlatformName();
public native String getDeviceName();
@Override
protected void onCreate(Bundle savedInstanceState) {
super.onCreate(savedInstanceState);
setContentView(R.layout.activity_main);
TextView testView2=(TextView)findViewById(R.id.textView2);
TextView testView4=(TextView)findViewById(R.id.textView4);
TextView testView6=(TextView)findViewById(R.id.textView6);
testView6.setText(testopencl());
testView2.setText(getPlatformName());
testView4.setText(getDeviceName());
}
}

6、NDK代码：ocl.cpp文件

[cpp]view plain copy

#include
#include
#include
#include
#include
#include"com_example_ocl_MainActivity.h"
#define LEN(arr) sizeof(arr) / sizeof(arr[0])
#define N 1024
#define NUM_THREAD 128
cl_uint num_device;
cl_uint num_platform;
cl_platform_id *platform;
cl_device_id *devices;
cl_int err;
cl_context context;
cl_command_queue cmdQueue;
cl_mem buffer,sum_buffer;
cl_program program ;
cl_kernel kernel;
const char* src[] = {
"__kernel void redution(\n"
"__global int *data,\n"
"__global int *output,\n"
"__local int *data_local\n"
")\n"
" {\n"
"int gid=get_group_id(0); \n"
"int tid=get_global_id(0); \n"
"int size=get_local_size(0); \n"
"int id=get_local_id(0); \n"
"data_local[id]=data[tid]; \n"
"barrier(CLK_LOCAL_MEM_FENCE); \n"
"for(int i=size/2; i>0; i>>=1){\n"
"if(id
"data_local[id]+=data_local[id+i]; \n"
"}\n"
"barrier(CLK_LOCAL_MEM_FENCE); \n"
"}\n"
"if(id==0){\n"
"output[gid]=data_local[0]; \n"
"}\n"
" }\n"
};
int num_block;
inttest()
{
int* in,*out;
num_block=N/NUM_THREAD;
in=(int*)malloc(sizeof(int)*N);
out=(int*)malloc(sizeof(int)*num_block);
for(int i=0; i
in[i]=1;
}
Init_OpenCL();
Context_cmd();
Create_Buffer(in);
Create_program();
Set_arg();
Execution();
CopyOutResult(out);
int sum=0;
for(int i=0; i
sum+=out[i];
}
return sum;
}
JNIEXPORT jstring JNICALL Java_com_example_ocl_MainActivity_testopencl (JNIEnv * env, jobject thisobject)
{
char result[10];
sprintf(result,"%d\n",test());
return env->NewStringUTF(result);
}
JNIEXPORT jstring JNICALL Java_com_example_ocl_MainActivity_getPlatformName(JNIEnv *env , jobject thisobject)
{
char buffer[1024];
clGetPlatformInfo(platform[0],CL_PLATFORM_NAME,sizeof(buffer),buffer,NULL);
return env->NewStringUTF(buffer);
}
JNIEXPORT jstring JNICALL Java_com_example_ocl_MainActivity_getDeviceName(JNIEnv *env , jobject thisobject)
{
char buffer[1024];
clGetDeviceInfo(devices[0],CL_DEVICE_NAME,sizeof(buffer),buffer,NULL);
return env->NewStringUTF(buffer);
}
void Init_OpenCL()
{
size_t nameLen1;
char platformName[1024];
err = clGetPlatformIDs(0, 0, &num_platform);
platform=(cl_platform_id*)malloc(sizeof(cl_platform_id)*num_platform);
err = clGetPlatformIDs(num_platform, platform, NULL);
err=clGetDeviceIDs(platform[0],CL_DEVICE_TYPE_GPU,0,NULL,&num_device);
devices=(cl_device_id*)malloc(sizeof(cl_device_id)*num_device);
err=clGetDeviceIDs(platform[0],CL_DEVICE_TYPE_GPU,num_device,devices,NULL);
}
void Context_cmd()
{
context=clCreateContext(NULL,num_device,devices,NULL,NULL,&err);
cmdQueue=clCreateCommandQueue(context,devices[0],0,&err);
}
void Create_Buffer(int *data)
{
buffer=clCreateBuffer(context,CL_MEM_READ_ONLY|CL_MEM_COPY_HOST_PTR,sizeof(int)*N,data,&err);
sum_buffer=clCreateBuffer(context,CL_MEM_WRITE_ONLY,sizeof(int)*num_block,0,&err);
}
void Create_program()
{
program=clCreateProgramWithSource(context, LEN(src), src, NULL, NULL);
err=clBuildProgram(program,num_device,devices,NULL,NULL,NULL);
kernel = clCreateKernel(program, "redution", NULL);
}
void Set_arg()
{
err=clSetKernelArg(kernel,0,sizeof(cl_mem),&buffer);
err=clSetKernelArg(kernel,1,sizeof(cl_mem),&sum_buffer);
err=clSetKernelArg(kernel,2,sizeof(int)*NUM_THREAD,NULL);
}
void Execution()
{
const size_t globalWorkSize[1]={N};
const size_t localWorkSize[1]={NUM_THREAD};
err=clEnqueueNDRangeKernel(cmdQueue,kernel,1,NULL,globalWorkSize,localWorkSize,0,NULL,NULL);
clFinish(cmdQueue);
}
void CopyOutResult(int*out)
{
err=clEnqueueReadBuffer(cmdQueue,sum_buffer,CL_TRUE,0,sizeof(int)*num_block,out,0,NULL,NULL);
}