8 Replies Latest reply on Dec 25, 2009 2:38 PM by spectral

    How to pass a lot of datas to openCL ?

    spectral

      Hi,

      In my application I have an OpenCL kernel that search for intersection between a ray and a set of triangles.

      I pass the list of triangles as a pointer, (I have up to 100.000 triangles now and can have more...)

      But, is there  a way to avoid to send all the triangles at each kernel call ?

      Theses datas are statics for a same object (I have several objects that contains a set of triangles).

       

      Thanks

        • How to pass a lot of datas to openCL ?
          n0thing

          You don't have to send the triangles to the device each time you are executing a kernel.

          Create a read-only buffer on device and enqueue write data to it only once, then you can re-use that buffer as your input triangle source as many times as you want.

          You can also use a constant buffer which may be cached.

            • How to pass a lot of datas to openCL ?
              spectral

              Thanks for your answer,

              And how can I :

              A) create a read-only buffer and access the read only buffer ?
              B) create a constant buffer and access the constant buffer ?

               

              Thx

                • How to pass a lot of datas to openCL ?
                  n0thing
                  To create a read-only buffer use flag : CL_MEM_READ_ONLY when using the function clCreateBuffer. See section 5.2.1 of OpenCL spec for more details. You can access buffers allocated inside global memory as constant buffers by using __constant as address space qualifier in your kernel arguments. For example: __kernel void sobel_filter(__constant uchar4* inputImage, __global uchar4* outputImage) The constant qualifier in the first argument to kernel is used to allocate 'inputImage' buffer in constant memory, this memory may be cached so you can get a good speedup if you read sequentially from this memory. See section 6.5 and in particular 6.5.3 for more details.