1 Reply Latest reply on Dec 3, 2018 1:39 AM by dipak

    How to pass a pointer to structure that contains a value in opencl

    avinashkrc

      I am trying to work on this code but everytime I gets runtime error "error: field may not be qualified with an address space global uint64_t external;

      Here is my data structure in hdr.h:

      typedef unsigned long uint64_t;
      typedef struct _elem 

        global uint64_t external;
      } elem; 

      typedef struct _buf

        global elem *kbuf; 
      }buf;

      My OpenCL kernel is:

      #include "hdr.h"
      __kernel void svmtest(global buf *tx) {
        if(get_global_id(0) == 0)
        printf((__constant char *)" %u %s\n", (tx[0].kbuf->external), "test");

      }

      My OpenCL C code in brief is as follows:

           parent =(elem*)clSVMAlloc(context, CL_MEM_READ_WRITE|CL_MEM_SVM_FINE_GRAIN_BUFFER , sizeof(elem),0);
           buf *p =(buf*)clSVMAlloc(context, CL_MEM_READ_WRITE|CL_MEM_SVM_FINE_GRAIN_BUFFER , sizeof(buf),0);

           parent->external = 9;

           p[0].kbuf = parent;

           status = clSetKernelArgSVMPointer(kernel, index++, p);
           status= clSetKernelExecInfo(
                kernel,
                CL_KERNEL_EXEC_INFO_SVM_PTRS,
                sizeof(parent),  &parent
            );

      However if I change uint64_ external to uint64_t * external in hdr.h and modify the code accordingly then everything works fine. Can't I pass values in structures, is it must to pass a pointer?

        • Re: How to pass a pointer to structure that contains a value in opencl
          dipak

          Only variables defined at program scope and static variables inside a function can be declared in the global address space. As the OpenCL spec says:

          __global (or global)

          --------------------------------

          The __global or global address space name is used to refer to memory objects (buffer or image objects) allocated from the global memory pool.

          A buffer memory object can be declared as a pointer to a scalar, vector or user-defined struct. This allows the kernel to read and/or write any location in the buffer.

          ..

          Some examples are:

          global float4 *color; // An array of float4 elements

          typedef struct {

          float a[3];

          int b[2];

          } foo_t;

          global foo_t *my_info; // An array of foo_t elements.

          ...

          Variables defined at program scope and static variables inside a function can also be declared in the global address space.

          ...

          For more details, please refer this: The OpenCL™ C 2.0 Specification

           

          Thanks.

          2 of 2 people found this helpful