高通平台下安卓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代码：

public class MainActivity extends ActionBarActivity {static    {        System.loadLibrary("ocl");    }public native String  testopencl();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文件

#include <jni.h>#include <CL/cl.h>#include<malloc.h>#include<stdio.h>#include<stdlib.h>#include"com_example_ocl_MainActivity.h"#define LEN(arr) sizeof(arr) / sizeof(arr[0])#define N 1024#define NUM_THREAD 128cl_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<i){   \n" "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;int  test(){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<N;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<num_block;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);}

7、Android.mk代码：

LOCAL_PATH := $(call my-dir)include $(CLEAR_VARS)LOCAL_MODULE    := oclLOCAL_SRC_FILES := ocl.cppLOCAL_LDFLAGS += -llog  -lOpenCLinclude $(BUILD_SHARED_LIBRARY)

8、执行结果：

搞定！！！