CLBufferRead

读取一个 OpenCL缓冲到一个数组并且返回读元件的数量.

uint  CLBufferRead(
   int          buffer,                    // 处理一个OpenCL缓冲
   const void&  data[],                    // 一个值的数组
   uint         buffer_offset=0,           // 在字节缓冲区偏移的OpenCL，默认为0
   uint         data_offset=0,             // 在元素的数组的偏移量，默认为0
   uint         data_count=WHOLE_ARRAY     // 用于读取的缓冲值的个数，默认为整个缓冲区
   );

这里还有处理矩阵和向量的版本。

读取OpenCL缓冲区到矩阵，如果成功则返回true。

uint  CLBufferRead(
   int           buffer,                    // OpenCL缓冲区的句柄
   uint          buffer_offset,             // 以字节为单位的OpenCL缓冲区抵消
   const matrxi& mat,                       // 用于接收缓冲区值的矩阵
   ulong         rows=-1,                   // 矩阵中的行数
   ulong         cols=-1                    // 矩阵中的列数
   );

读取OpenCL缓冲区到向量，如果成功则返回true。

uint  CLBufferRead(
   int           buffer,                    // OpenCL缓冲区的句柄
   uint          buffer_offset,             // 以字节为单位的OpenCL缓冲区抵消
   const vector& vec,                       // 用于接收缓冲区值的向量
   ulong         size-1,                    // 向量长度
   );

参数

buffer: [in] 一个OpenCL缓冲控制.
data[]: [in] 一个用于从OpenCL的缓冲区接收的值的数组。通过引用传递.
buffer_offset: [in] 一个在字节区偏移的OpenCL缓冲，从读取开始处。默认情况下，读取开始于正开头的缓冲.
data_offset: [in] 用于书写的OpenCL缓冲的值的第一个数组元素的索引。默认情况下，读写的值为数组从零开始指数.
data_count: [in] 值的数目应被读取。默认读取整个OpenCL缓冲区.
mat: [out] 用于将数据写入缓冲区的矩阵可以是三种类型中的任何一种 — matrix、matrixf或matrixc.
vec: [out] 用于将数据写入缓冲区的向量可以是三种类型中的任何一种 — vector、vectorf或vectorc.
rows=-1: [in] 如果该参数被指定，也应该指定cols参数。如果没有指定新矩阵维度，则使用之前的矩阵维度。如果该值为 -1，则行数不变.
cols=-1: [in] 如果该参数被指定，也应该跳过rows参数。矩阵遵循以下规则：要么两个参数都指定，要么都不指定，否则将发生错误。如果两个参数(rows和cols)都指定，则指定矩阵大小。-1情况下，则列数不变.
size=-1: [in] 如果没有指定该参数或参数值为-1，则向量长度不变.

返回值

读取元素个数。错误的情况下返回0。有关错误信息，使用 GetLastError() 函数.

如果矩阵或向量成功处理，返回true，否则返回false.

注意

对于一维数组，考虑到 AS_SERIES 旗帜，元素的数量，从写数据到一个OpenCL缓冲开始计算.

两个或两个以上的尺寸数组是一维的。在这种情况下，data_offset 是要在演示文稿中跳过的元素个数，而不是在第一维的元素个数.

使用方程式计算Pi的示例：

pi_calculation_formula

#define  _num_steps        1000000000
#define  _divisor          40000
#define  _step             1.0 / _num_steps
#define  _intrnCnt         _num_steps / _divisor

//+------------------------------------------------------------------+
//|                                                                  |
//+------------------------------------------------------------------+
string D2S(double arg, int digits) { return DoubleToString(arg, digits); }
string I2S(int arg)                { return IntegerToString(arg); }

//--- OpenCL程序代码
const string clSource=
  "#define _step "+D2S(_step, 12)+"                   \r\n"
  "#define _intrnCnt "+I2S(_intrnCnt)+"               \r\n"
  "                                                   \r\n"
  "__kernel void Pi( __global double *out )           \r\n"
  "{                                                  \r\n"
  "  int i = get_global_id( 0 );                      \r\n"
  "  double partsum = 0.0;                            \r\n"
  "  double x = 0.0;                                  \r\n"
  "  long from = i * _intrnCnt;                       \r\n"
  "  long to = from + _intrnCnt;                      \r\n"
  "  for( long j = from; j < to; j ++ )               \r\n"
  "  {                                                \r\n"
  "     x = ( j + 0.5 ) * _step;                      \r\n"
  "     partsum += 4.0 / ( 1. + x * x );              \r\n"
  "  }                                                \r\n"
  "  out[ i ] = partsum;                              \r\n"
  "}                                                  \r\n";

//+------------------------------------------------------------------+
//| 脚本程序起始函数                                                   |
//+------------------------------------------------------------------+
int OnStart()
 {
  Print("Pi Calculation: step = "+D2S(_step, 12)+"; _intrnCnt = "+I2S(_intrnCnt));
//--- 比较OpenCL环境
  int clCtx;
  if((clCtx=CLContextCreate(CL_USE_GPU_ONLY))==INVALID_HANDLE)
   {
    Print("OpenCL not found");
    return(-1);
   }
  int clPrg = CLProgramCreate(clCtx, clSource);
  int clKrn = CLKernelCreate(clPrg, "Pi");
  int clMem=CLBufferCreate(clCtx, _divisor*sizeof(double), CL_MEM_READ_WRITE);
  CLSetKernelArgMem(clKrn, 0, clMem);

  const uint offs[1]  = {0};
  const uint works[1] = {_divisor};
//--- 启动OpenCL程序
  ulong start=GetMicrosecondCount();
  if(!CLExecute(clKrn, 1, offs, works))
   {
    Print("CLExecute(clKrn, 1, offs, works) failed! Error ", GetLastError());
    CLFreeAll(clMem, clKrn, clPrg, clCtx);
    return(-1);
   }
//--- 从OpenCL设备获取结果
  vector buffer(_divisor);
  if(!CLBufferRead(clMem, 0, buffer))
   {
    Print("CLBufferRead(clMem, 0, buffer) failed! Error ", GetLastError());
    CLFreeAll(clMem, clKrn, clPrg, clCtx);
    return(-1);
   }
//--- 将所有值加起来计算Pi
  double Pi=buffer.Sum()*_step;

  double time=(GetMicrosecondCount()-start)/1000.;
  Print("OpenCL: Pi calculated for "+D2S(time, 2)+" ms");
  Print("Pi = "+DoubleToString(Pi, 12));
//--- 空闲内存
  CLFreeAll(clMem, clKrn, clPrg, clCtx);
//--- 成功
  return(0);
 }
  /*
  Pi Calculation: step = 0.000000001000; _intrnCnt = 25000
  OpenCL: GPU device 'Ellesmere' selected
  OpenCL: Pi calculated for 99.98 ms
  Pi = 3.141592653590
  */
//+------------------------------------------------------------------+
//| 空闲内存的辅助程序                                                  |
//+------------------------------------------------------------------+
void CLFreeAll(const int clMem, const int clKrn, const int clPrg, const int clCtx)
 {
  CLBufferFree(clMem);
  CLKernelFree(clKrn);
  CLProgramFree(clPrg);
  CLContextFree(clCtx);
 }

CLBufferWrite CLExecute