Forum Discussion

Altera_Forum's avatar
Altera_Forum
Icon for Honored Contributor rankHonored Contributor
9 years ago

How can I use local memory.

I read Optimizing Local Memory Accesses of Altera SDK for OpenCL Best Prasctices Guides.

and I can see that I can use local memory used by clSetKernelArg.

so I write my source as below

/////////kernel source////////

__kernel void test_kernel(__attribute __((local_mem_size(2048)))__local int const * restrict table)

{

...

}

////////////////////////////////

/////////my source////////

status = clSetKernelArg(kernel, argi++, 256*sizeof(cl_mem), &table);

/////////////////////////////

1) I compile my kernel and I see warning massage as bellow.

Compiler Warning: Pointer to local memory argument with no store to it.

2)and execute my program. I can see error massage as bellow.

ERROR: CL_INVALID_ARG_VALUE

How can I use local memory used clSetKernelArg????

thanks.

5 Replies

  • Altera_Forum's avatar
    Altera_Forum
    Icon for Honored Contributor rankHonored Contributor
    9 years ago

    Hi there,

    You are not using clSetKernelArg correctly as the proper way to set a pointer to local memory is:

    status = clSetKernelArg(kernel, argi++, sizeof(<type>)*<size_of_array>, null);

    so for example, if you want to set an array of chars of size 256, you would have to write:

    status = clsetkernelarg(kernel, argi++, sizeof(char)*256, NULL);

    That's all. Good luck!
  • Altera_Forum's avatar
    Altera_Forum
    Icon for Honored Contributor rankHonored Contributor
    9 years ago

    But what about copying the value of table into the kernel?

    thanks.

    --- quote start ---

    hi there,

    you are not using clsetkernelarg correctly as the proper way to set a pointer to local memory is:

    status = clsetkernelarg(kernel, argi++, sizeof(<type>)*<size_of_array>, NULL);

    So for example, if you want to set an array of chars of size 256, you would have to write:

    status = clSetKernelArg(kernel[i], argi++, sizeof(char)*256, NULL);

    That's all. Good luck!

    --- Quote End ---

  • Altera_Forum's avatar
    Altera_Forum
    Icon for Honored Contributor rankHonored Contributor
    9 years ago

    --- Quote Start ---

    I read Optimizing Local Memory Accesses of Altera SDK for OpenCL Best Prasctices Guides.

    and I can see that I can use local memory used by clSetKernelArg.

    so I write my source as below

    /////////kernel source////////

    __kernel void test_kernel(__attribute __((local_mem_size(2048)))__local int const * restrict table)

    {

    ...

    }

    ////////////////////////////////

    /////////my source////////

    status = clSetKernelArg(kernel, argi++, 256*sizeof(cl_mem), &table);

    /////////////////////////////

    1) I compile my kernel and I see warning massage as bellow.

    Compiler Warning: Pointer to local memory argument with no store to it.

    2)and execute my program. I can see error massage as bellow.

    ERROR: CL_INVALID_ARG_VALUE

    How can I use local memory used clSetKernelArg????

    thanks.

    --- Quote End ---

    It's not possible to transfer data from the host to local memory. You have to transfer it to global memory and have the kernel load the data into local memory.

    The only advantage to declaring local memory in a kernel parameter is that you get to choose the size at runtime instead of kernel compile time.
  • TJone2's avatar
    TJone2
    Icon for New Contributor rankNew Contributor
    7 years ago

    What is irritating about this thread is the OP is correct that having __local memory pointers on the kernel interface makes no sense since the host cannot stuff data into local memory under the Altera/Intel regime. But no where is this said and there is even and example of a kernel with __local pointer arguments in the programming spec.

  • HRZ's avatar
    HRZ
    Icon for Frequent Contributor rankFrequent Contributor
    7 years ago

    @TJone2, this has nothing to do with Intel/Altera; you cannot transfer data directly from the host to the device local memory on any hardware, regardless of what OpenCL SDK or hardware you use, simply because device local memory is not addressable from the host side. In the case of FPGAs, since the size of local buffers has to be determined at compile-time anyway, there is really no point in defining local buffers in the kernel header. Intel FPGA SDK for OpenCL has to support this type of definition since it is part of the standard; however, I would personally avoid this method at all costs since it creates unnecessary limitations (e.g. buffer size has to be a power of two, a limitation that does not exist if the local buffer is defined inside of the kernel).