UPGRADE YOUR BROWSER

We have detected your current browser version is not the latest one. Xilinx.com uses the latest web technologies to bring you the best online experience possible. Please upgrade to a Xilinx.com supported browser:Chrome, Firefox, Internet Explorer 11, Safari. Thank you!

cancel
Showing results for 
Search instead for 
Did you mean: 
Highlighted
Observer kaliuday
Observer
7,759 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
1 Solution

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

Re: SDAccel OpenCL Pipeline

Jump to solution

Hi @herver

 

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

 

 

0 Kudos
4 Replies
Xilinx Employee
Xilinx Employee
7,736 Views
Registered: ‎08-17-2011

Re: SDAccel OpenCL Pipeline

Jump to solution

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 kaliuday
Observer
7,728 Views
Registered: ‎04-02-2015

Re: SDAccel OpenCL Pipeline

Jump to solution
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
Xilinx Employee
Xilinx Employee
7,684 Views
Registered: ‎08-17-2011

Re: SDAccel OpenCL Pipeline

Jump to solution

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 kaliuday
Observer
14,368 Views
Registered: ‎04-02-2015

Re: SDAccel OpenCL Pipeline

Jump to solution

Hi @herver

 

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

 

 

0 Kudos