cancel
Showing results for 
Show  only  | Search instead for 
Did you mean: 
Observer
Observer
8,188 Views
Registered: ‎04-02-2015

SDAccel OpenCL Pipeline

Jump to solution
    • Is it possible to pipeline entire body of the function using OpenCL pipeline attribute ? or is there any other alternative to achieve the same? (It works in HLS kernel by using #pragma HLS PIPELINE).

Sample HLS Code Snippet :

 

void dummy(short r,short c,pe *p, pe*ppx, uint2_t *d, uint2_t *q){
#pragma HLS PIPELINE
    short nw, w, n;
 
    if (r == 0){
        n = 0;
        nw = 0;
    }else{
        n = px->p;
     }
    w = 123;
    uint2_t d1 = d[c];
    uint2_t q1 = q[r];
    update_px(px, d1, q1, n, nw, w, r, c);
}
Tags (2)
0 Kudos
Reply
1 Solution

Accepted Solutions
Observer
Observer
14,796 Views
Registered: ‎04-02-2015

Hi @herver

 

Thanks a lot for your advice. It worked as expected. 

 

 

View solution in original post

0 Kudos
Reply
4 Replies
Xilinx Employee
Xilinx Employee
8,165 Views
Registered: ‎08-17-2011

hello @kaliuday

 

xcl_pipeline_workitems is the example, note that you need to create a region - the attribute apply to a region!

this is one of the templtes that we have in the tools.

 

kernel __attribute__ ((reqd_work_group_size(3,1,1)))
void mykernel(...)
{
    __attribute__((xcl_pipeline_workitems)) {
        int tid = get_global_id(0);
        op_Read(tid);
        op_Compute(tid);
        op_Write(tid);
    }
}

- Hervé

SIGNATURE:
* New Dedicated Vivado HLS forums* http://forums.xilinx.com/t5/High-Level-Synthesis-HLS/bd-p/hls
* Readme/Guidance* http://forums.xilinx.com/t5/New-Users-Forum/README-first-Help-for-new-users/td-p/219369

* Please mark the Answer as "Accept as solution" if information provided is helpful.
* Give Kudos to a post which you think is helpful and reply oriented.
Observer
Observer
8,157 Views
Registered: ‎04-02-2015
I am aware of this attribute, but intuition was like "xcl_pipeline_wokitems" if and only if we have multiple threads launched and get_global_id() used to index. will it work if I use this attribute in the code snippet which I posted in this chain ? my kernel runs with single thread of execution.

Thanks for your prompt reply @herver
0 Kudos
Reply
Xilinx Employee
Xilinx Employee
8,113 Views
Registered: ‎08-17-2011

HI @kaliuday

 

SDAccel takes C / C ++ / OpenCL kernels so if you feel like you are more confortable with C / C++ , then use that and the VHLS pragmas - make sure to read the UG to respect the function signature needed.

The xcl pipeline workitems should have same effect as the pragma pipeline that you show.

With it, all subloops are unrolled and the datapath is pipelined so that it can process data at II=1

In VHLS if you have no loop that's the same as in OpenCL kernel with a single thread.

 

I hope this helps.

- Hervé

SIGNATURE:
* New Dedicated Vivado HLS forums* http://forums.xilinx.com/t5/High-Level-Synthesis-HLS/bd-p/hls
* Readme/Guidance* http://forums.xilinx.com/t5/New-Users-Forum/README-first-Help-for-new-users/td-p/219369

* Please mark the Answer as "Accept as solution" if information provided is helpful.
* Give Kudos to a post which you think is helpful and reply oriented.
Observer
Observer
14,797 Views
Registered: ‎04-02-2015

Hi @herver

 

Thanks a lot for your advice. It worked as expected. 

 

 

View solution in original post

0 Kudos
Reply