Altera_Forum
Honored Contributor
10 years agorestrict Keyword & Pointer Aliasing
Hello, I have read over the Altera Optimization Guide and Programming Guide concerning the restrict keyword:
"Insert the restrict keyword in pointer arguments whenever possible. Including the restrict keyword in pointer arguments prevents the Altera Offline Compiler (AOC) from creating unnecessary memory dependencies between non-conflicting load and store operations." Based on that, the restrict keyword tells the compiler that you promise not to alias pointer A and B in the following situation: __kernel void myKernel (__global int * restrict PtrA, __global int * restrict PtrB) Here both A and B point to global memory, but the programmer is telling the compiler that PtrA and PtrB will never point to the same address. My problem is that I have two pointers that are derived from the same kernel input pointer, similar too: __kernel void myKernel (__local int * restrict ptrA, int Offset) { for( int i = 0; i < 4; i++ ) ptrA[i] = ptrA[i+Offset]; } The value of Offset that I pass will guarantee that the read and write will never overlap, but when I compile the kernel, I get a report message like: | Memory dependency on Load Operation from: | [1]:4 | | Store Operation | [1]:4 | So even though I have used the restrict keyword, aoc is still seeing a dependency between the read and write. How can I tell aoc to ignore the dependency in this case?