CLBufferWrite

uint  CLBufferWrite(
   int          buffer,                    // 处理一个OpenCL缓冲
   const void&  data[],                    // 值的数组
   uint         buffer_offset=0,           // 字节区偏移的OpenCL缓冲，默认为0
   uint         data_offset=0,             // 元素中数组的偏移，默认为0
   uint         data_count=WHOLE_ARRAY     // 从阵列写入的值的数量，默认整个阵列
   );

uint  CLBufferWrite(
   int           buffer,                    // OpenCL缓冲区的句柄
   uint          buffer_offset,             // 以字节为单位的OpenCL缓冲区抵消
   matrix<T>     &mat                       // 用于写入缓冲区的值矩阵
   );

uint  CLBufferWrite(
   int           buffer,                    // OpenCL缓冲区的句柄
   uint          buffer_offset,             // 以字节为单位的OpenCL缓冲区抵消
   vector<T>     &vec                       // 用于写入缓冲区的值向量
   );

#define M       3000      // 第一个矩阵中的行数
#define K       2000      // 第一个矩阵中的列数等于第二个矩阵的行数
#define N       3000      // 第二个矩阵中的列数
 
//+------------------------------------------------------------------+
const string clSrc=
  "#define N     "+IntegerToString(N)+"                              \r\n"
  "#define K     "+IntegerToString(K)+"                              \r\n"
  "                                                                  \r\n"
  "__kernel void matricesMul( __global float *in1,                   \r\n"
  "                           __global float *in2,                   \r\n"
  "                           __global float *out  )                 \r\n"
  "{                                                                 \r\n"
  "  int m = get_global_id( 0 );                                     \r\n"
  "  int n = get_global_id( 1 );                                     \r\n"
  "  float sum = 0.0;                                                \r\n"
  "  for( int k = 0; k < K; k ++ )                                   \r\n"
  "     sum += in1[ m * K + k ] * in2[ k * N + n ];                  \r\n"
  "  out[ m * N + n ] = sum;                                         \r\n"
  "}                                                                 \r\n";
//+------------------------------------------------------------------+
//| 脚本程序起始函数                                                   |
//+------------------------------------------------------------------+
void OnStart()
 {
//--- 初始化随机数生成器
  MathSrand((int)TimeCurrent());
//--- 用随机值填写给定大小的矩阵
  matrixf mat1(M, K, MatrixRandom) ;    // 第一个矩阵
  matrixf mat2(K, N, MatrixRandom);     //第二个矩阵
 
//--- 使用原始方法计算矩阵的乘积
  uint start=GetTickCount();
  matrixf matrix_naive=matrixf::Zeros(M, N);// 在这里设置两个矩阵相乘的结果
  for(int m=0; m<M; m++)
    for(int k=0; k<K; k++)
      for(int n=0; n<N; n++)
        matrix_naive[m][n]+=mat1[m][k]*mat2[k][n];
  uint time_naive=GetTickCount()-start;   
     
//--- 通过MatMull计算矩阵的乘积
  start=GetTickCount();
  matrixf matrix_matmul=mat1.MatMul(mat2);
  uint time_matmul=GetTickCount()-start;     
  
//--- 在OpenCL中计算矩阵的乘积
  matrixf matrix_opencl=matrixf::Zeros(M, N);
  int cl_ctx;             // 环境句柄
  if((cl_ctx=CLContextCreate(CL_USE_GPU_ONLY))==INVALID_HANDLE)
   {
    Print("OpenCL not found, leaving");
    return;
   }
  int cl_prg;             // 程序句柄 
  int cl_krn;             // 内核句柄
  int cl_mem_in1;         // 第一个（输入）缓冲区句柄
  int cl_mem_in2;         // 第二个（输入）缓冲区句柄
  int cl_mem_out;         // 第三个（输出）缓冲区句柄
//--- 创建程序和内核
  cl_prg = CLProgramCreate(cl_ctx, clSrc);
  cl_krn = CLKernelCreate(cl_prg, "matricesMul");
//--- 为三个矩阵创建所有三个缓冲区
  cl_mem_in1=CLBufferCreate(cl_ctx, M*K*sizeof(float), CL_MEM_READ_WRITE);
  cl_mem_in2=CLBufferCreate(cl_ctx, K*N*sizeof(float), CL_MEM_READ_WRITE);
//--- 第三个矩阵 - 输出
  cl_mem_out=CLBufferCreate(cl_ctx, M*N*sizeof(float), CL_MEM_READ_WRITE);
//--- 设置内核参数
  CLSetKernelArgMem(cl_krn, 0, cl_mem_in1);
  CLSetKernelArgMem(cl_krn, 1, cl_mem_in2);
  CLSetKernelArgMem(cl_krn, 2, cl_mem_out);
//--- 将矩阵写入设备缓冲区
  CLBufferWrite(cl_mem_in1, 0, mat1);
  CLBufferWrite(cl_mem_in2, 0, mat2);
  CLBufferWrite(cl_mem_out, 0, matrix_opencl);
//--- OpenCL代码执行时间开始
  start=GetTickCount();
//--- 设置任务工作区参数，并执行OpenCL程序
  uint  offs[2] = {0, 0};
  uint works[2] = {M, N};
  start=GetTickCount();  
  bool ex=CLExecute(cl_krn, 2, offs, works);
//--- 计算矩阵结果
  if(CLBufferRead(cl_mem_out, 0, matrix_opencl))
    PrintFormat("[%d x %d] matrix read: ", matrix_opencl.Rows(), matrix_opencl.Cols());
   else
      Print("CLBufferRead(cl_mem_out, 0, matrix_opencl failed. Error ",GetLastError()); 
  uint time_opencl=GetTickCount()-start;   
  Print("Compare calculation time using each method");
  PrintFormat("Naive product time = %d ms",time_naive);
  PrintFormat("MatMul product time = %d ms",time_matmul);
  PrintFormat("OpenCl product time = %d ms",time_opencl);  
//--- 释放所有OpenCL环境
  CLFreeAll(cl_ctx, cl_prg, cl_krn, cl_mem_in1, cl_mem_in2, cl_mem_out);
 
//--- 比较所有获得的结果矩阵
  Print("How many discrepancy errors are there between result matrices?");
  ulong errors=matrix_naive.Compare(matrix_matmul,(float)1e-12);
  Print("matrix_direct.Compare(matrix_matmul,1e-12)=",errors);
  errors=matrix_matmul.Compare(matrix_opencl,float(1e-12));
  Print("matrix_matmul.Compare(matrix_opencl,1e-12)=",errors);
/*
  结果：
   
   [3000 x 3000] matrix read: 
  比较每种方法的计算时间
   Naive product time = 54750 ms
   MatMul product time = 4578 ms
   OpenCl product time = 922 ms
  结果矩阵之间有多少误差？
   matrix_direct.Compare(matrix_matmul,1e-12)=0
   matrix_matmul.Compare(matrix_opencl,1e-12)=0
*/  
 }
//+------------------------------------------------------------------+
//| 用随机值填写矩阵                                                   |
//+------------------------------------------------------------------+
void MatrixRandom(matrixf& m)
 {
  for(ulong r=0; r<m.Rows(); r++)
   {
    for(ulong c=0; c<m.Cols(); c++)
     {
      m[r][c]=(float)((MathRand()-16383.5)/32767.);
     }
   }
 }
//+------------------------------------------------------------------+
//| 释放所有OpenCL环境                                                |
//+------------------------------------------------------------------+
void CLFreeAll(int cl_ctx, int cl_prg, int cl_krn,
               int cl_mem_in1, int cl_mem_in2, int cl_mem_out)
 {
//--- 以相反顺序删除OpenCL创建的所有环境
  CLBufferFree(cl_mem_in1);
  CLBufferFree(cl_mem_in2);
  CLBufferFree(cl_mem_out);
  CLKernelFree(cl_krn);
  CLProgramFree(cl_prg);
  CLContextFree(cl_ctx);
 }