Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Is barrier in kernel supported or going to be supported? #180

Open
fxzjshm opened this issue May 4, 2022 · 2 comments
Open

Is barrier in kernel supported or going to be supported? #180

fxzjshm opened this issue May 4, 2022 · 2 comments
Labels
enhancement New feature or request

Comments

@fxzjshm
Copy link

fxzjshm commented May 4, 2022

barrier() of sycl::nd_item seems not working, the log says:

===>The following messages were generated while  performing high-level synthesis for kernel: geIXT1_EEERKT0_EUlvE_85uG06fiV Log file: /tmp/dedisp-1337dez7lrhv1u/vxx_comp_tmp/geIXT1_EEERKT0_EUlvE_85uG06fiV/geIXT1_EEERKT0_EUlvE_85uG06fiV/vitis_hls.log :
ERROR: [v++ 214-194] in function 'geIXT1_EEERKT0_EUlvE_85uG06fiV': Undefined function ControlBarrier
ERROR: [v++ 214-135] Syn check fail!
ERROR: [v++ 200-1715] Encountered problem during source synthesis
ERROR: [v++ 60-300] Failed to build kernel(ip) geIXT1_EEERKT0_EUlvE_85uG06fiV, see log for details: /tmp/dedisp-1337dez7lrhv1u/vxx_comp_tmp/geIXT1_EEERKT0_EUlvE_85uG06fiV/geIXT1_EEERKT0_EUlvE_85uG06fiV/vitis_hls.log
ERROR: [v++ 60-773] In '/tmp/dedisp-1337dez7lrhv1u/vxx_comp_tmp/geIXT1_EEERKT0_EUlvE_85uG06fiV/geIXT1_EEERKT0_EUlvE_85uG06fiV/vitis_hls.log', caught Tcl error: ERROR: [HLS 214-194] in function 'geIXT1_EEERKT0_EUlvE_85uG06fiV': Undefined function ControlBarrier
ERROR: [v++ 60-773] In '/tmp/dedisp-1337dez7lrhv1u/vxx_comp_tmp/geIXT1_EEERKT0_EUlvE_85uG06fiV/geIXT1_EEERKT0_EUlvE_85uG06fiV/vitis_hls.log', caught Tcl error: ERROR: [HLS 214-135] Syn check fail!
ERROR: [v++ 60-773] In '/tmp/dedisp-1337dez7lrhv1u/vxx_comp_tmp/geIXT1_EEERKT0_EUlvE_85uG06fiV/geIXT1_EEERKT0_EUlvE_85uG06fiV/vitis_hls.log', caught Tcl error: ERROR: [HLS 200-1715] Encountered problem during source synthesis
ERROR: [v++ 60-599] Kernel compilation failed to complete
ERROR: [v++ 60-592] Failed to finish compilation

Environment: Vitis 2021.2 & 2022.1, compiling with -fsycl-targets=fpga64_hls_hw_emu

@fxzjshm fxzjshm added the enhancement New feature or request label May 4, 2022
@keryell
Copy link
Member

keryell commented May 4, 2022

It used to work with the OpenCL/SPIR back-end since it was implemented by Vitis OpenCL but we deprecated it and we have moved to HLS back-end which is really single-task oriented, more suitable for FPGA.
The problem with parallel_for and barriers is that at some point you need a GPU thread emulator, requiring some memory to store the thread states and emulate the context switches. We could implement the compiler transformations to do it, but at the end this kind of very GPU-friendly code will probably work well on GPU but not with good performance on FPGA.
So this is not our priority, in comparison to implementing features really suitable for FPGA.
But what is the kind of applications are trying to run on FPGA?

@fxzjshm
Copy link
Author

fxzjshm commented May 5, 2022

Well, I know barriers and mem_fences isn't designed for FPGA, so I'm just asking for a possibility.
I was porting some CUDA-based scientific research program into SYCL (mainly on GPU and CPU), and my teacher let me try if it can be ported to FPGA as well. I will re-design it for FPGA then (well, if a can).
Thank for explanation!

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
enhancement New feature or request
Projects
None yet
Development

No branches or pull requests

2 participants