0
votes

I am trying to access the arrays, delivered via a call-signature into the system invoked OnCalculation() event-handler.

This the way it is written:

    int OnCalculate(const int       rates_total,
                    const int       prev_calculated,
                    const datetime &time[],
                    const double   &open[],
                    const double   &high[],
                    const double   &low[],
                    const double   &close[],
                    const long     &tick_volume[],
                    const long     &volume[],
                    const int      &spread[]
                    )
    {
    /* The rest code is written here
       ...
    */
    }

I am trying to merge the code with the OpenCL functions so that the program uses GPU for the tremendous calculations. But the issue is when I am trying to pass the values from OnCalculation() to the kernel for execution, I am getting error. See the following code is written inside OnCalculation()

  CLSetKernelArg( cl_krn, 0, start );
  CLSetKernelArg( cl_krn, 1, rates_total );
  CLSetKernelArg( cl_krn, 2, time );
  CLSetKernelArg( cl_krn, 3, high );
  CLSetKernelArg( cl_krn, 4, low );

Getting the following error:

'time' - invalid array access ADX.mq5 285 31
'high' - invalid array access ADX.mq5 286 31
'low' - invalid array access ADX.mq5 287 31

I don't know why is this problem happening. I am not able to pass the arrays from the OnCalculation().

Kindly, help me what I can do?

1

1 Answers

1
votes

It is impossible to just reference an MQL5 array[] object here

OpenCL starts a completely new code-execution eco-system, and MQL5-side data has to get "transferred" correctly there and back...


Using a mock-up trivial GPU-kernel that doubles an array received:

const string                                                           //  by default some GPU doesn't support doubles
cl_SOURCE = "#pragma OPENCL EXTENSION cl_khr_fp64 : enable      \r\n" //   cl_khr_fp64 directive is used to enable work with doubles
            "                                                   \r\n"
            "__kernel void Test_GPU( __global double *data,     \r\n" // [0]____GPU-kernel-side_CALL-SIGNATURE
            "                          const  int     N,        \r\n" // [1]____GPU-kernel-side_CALL-SIGNATURE
            "                          const  int     N_arrays  \r\n" // [2]____GPU-kernel-side_CALL-SIGNATURE
            "                          )                        \r\n"
            "{                                                  \r\n"
            "   uint kernel_index = get_global_id( 0 );         \r\n"
            "   if ( kernel_index > N_arrays ) return;          \r\n"
            "                                                   \r\n"
            "   uint local_start_offset = kernel_index * N;     \r\n"
            "   for (  int i = 0; i <  N; i++ )                 \r\n"
            "         data[i+local_start_offset] *= 2.0;        \r\n"
            "}                                                  \r\n";

// AFTER FIRST TESTING THE OpenCL DEVICES & THEIR CAPABILITIES ... ( see prev. posts )

#define ARRAY_SIZE   100  // size of the array
#define TOTAL_ARRAYS 5    // total arrays

// ONE CAN:
//---  SET OpenCL-specific handles' holders
int   cl_CONTEXT,  // an OpenCL-Context handle
      cl_PROGRAM,  // an OpenCL-Program handle
      cl_KERNEL,   // an OpenCL Device-Kernel handle
      cl_BUFFER;   // an OpenCL-buffer handle

uint  cl_offset[] = { 0 };            //--- prepare CLExecute() params
uint  cl_work[]   = { TOTAL_ARRAYS }; //--- global work size

double       DataArray2[];            //--- global mapping-object for data aimed to reach the GPU
ArrayResize( DataArray2,              //--- size it to fit data in
             ARRAY_SIZE * TOTAL_ARRAYS
             );
for ( int j = 0; j <  TOTAL_ARRAYS; j++ ) //--- fill mapped-arrays with data
{    uint local_offset = j * ARRAY_SIZE;               //--- set local start offset for j-th array
     for (  int i = 0; i <  ARRAY_SIZE; i++ )          //--- for j-th array
            DataArray2[i+local_offset] = MathCos(i+j); //--- fill array with some data
}

The principal structure of MQL5 / OpenCL setup is similar to this:

//--- INIT OpenCL 
   if ( INVALID_HANDLE == ( cl_CONTEXT = CLContextCreate() ) )
   {                        Print( "EXC: CLContextCreate() error = ", GetLastError() );
                            return( 1 ); // ---------------^ EXC/RET
   }
//--- NEXT create OpenCL program
   if ( INVALID_HANDLE == ( cl_PROGRAM = CLProgramCreate( cl_CONTEXT,
                                                          cl_SOURCE
                                                          )
                            )
        )
   {                        Print( "EXC: CLProgrameCreate() error = ", GetLastError() );
                            CLContextFree( cl_CONTEXT );
                            return( 1 ); // ----------------^ EXC/RET
   }
//--- NEXT create OpenCL kernel
   if ( INVALID_HANDLE == ( cl_KERNEL = CLKernelCreate( cl_PROGRAM,
                                                       "Test_GPU"
                                                        )
                            )
        )
   {                        Print( "EXC: CLKernelCreate() error = ", GetLastError() );
                            CLProgramFree( cl_PROGRAM );
                            CLContextFree( cl_CONTEXT ); 
                            return( 1 ); // --------------^ EXC/RET 
   }
//---  TRY: create an OpenCL cl_BUFFER object mapping
   if (  INVALID_HANDLE == ( cl_BUFFER = CLBufferCreate( cl_CONTEXT,
                                                         (uint) ( ARRAY_SIZE * TOTAL_ARRAYS * sizeof( double ),
                                                         CL_MEM_READ_WRITE
                                                         )
                             )
         )
   {                         Print( "EXC: CLBufferCreate() error == ", GetLastError() );
                             CLKernelFree(  cl_KERNEL  );
                             CLProgramFree( cl_PROGRAM );
                             CLContextFree( cl_CONTEXT ); 
                             return(1); // ----------------^ EXC/RET
   }
//--- NEXT: set OpenCL cl_KERNEL GPU-side-kernel call-parameters
   CLSetKernelArgMem( cl_KERNEL, 0, cl_BUFFER    ); // [0]____GPU-kernel-side_CALL-SIGNATURE
   CLSetKernelArg(    cl_KERNEL, 1, ARRAY_SIZE   ); // [1]____GPU-kernel-side_CALL-SIGNATURE
   CLSetKernelArg(    cl_KERNEL, 2, TOTAL_ARRAYS ); // [2]____GPU-kernel-side_CALL-SIGNATURE

//--- NEXT: write data into to OpenCL cl_BUFFER mapping-object
   CLBufferWrite( cl_BUFFER,
                  DataArray2
                  );

//---  MAY execute OpenCL kernel
   CLExecute( cl_KERNEL, 1, cl_offset, cl_work );

//---  MAY read data back, from OpenCL cl_BUFFER mapping-object
   CLBufferRead(  cl_BUFFER, DataArray2 );

   CLBufferFree(  cl_BUFFER  ); //--- FINALLY free OpenCL buffer cl_BUFFER mapping-object
   CLKernelFree(  cl_KERNEL  ); //--- FINALLY free OpenCL kernel object
   CLProgramFree( cl_PROGRAM ); //--- FINALLY free OpenCL programme object / handle
   CLContextFree( cl_CONTEXT ); //--- FINALLY free OpenCL cl_CONTEXT object / handle