Hello,
I am a student currently working on a bachelor project which consists of
studying the OpenCL standard, an OpenCL implementation (I chose POCL) and in
the end if I have time, making POCL support an accelerator developed at the
REDS institute of the university of applied sciences of Yverdon-Les-Bains in
Switzerland.
As an exercise/pretext for studying the internals of POCL, I developed a simple
virtual device that is intended to execute an OpenCL kernel and to be supported
by POCL. The underlying hardware of this device is the CPU, but, it has to have
a different architecture than the host's so that I am forced to study the whole
compilation process of an OCL kernel. In my case, I run POCL on an x86_64 host
and I want to run my device on an ARM system running Linux. So far, my
(virtual) device is registered by POCL and can run on an ARM system and make
data transfers with POCL. But I still can't generate the final binary, that is,
the shared object that is supposed to run on the device.
The problem I have right now is that the code generation in pocl_llvm_codegen,
the function that takes the optimized IR parallel.bc and generates the final
kernel.so, crashes with a segfault. In fact, the segfault appears in LLVM as
seen below in gdb :
Thread 1 "vecadd" received signal SIGSEGV, Segmentation fault.
0x00007ffff1f7ad1a in llvm::SelectionDAG::createOperands(llvm::SDNode*,
llvm::ArrayRef<llvm::SDValue>) () from /usr/lib/llvm-6.0/lib/libLLVM-6.0.so.1
(gdb) where
#0 0x00007ffff1f7ad1a in llvm::SelectionDAG::createOperands(llvm::SDNode*,
llvm::ArrayRef<llvm::SDValue>) () from /usr/lib/llvm-6.0/lib/libLLVM-6.0.so.1
#1 0x00007ffff1f885c4 in llvm::SelectionDAG::getNode(unsigned int, llvm::SDLoc
const&, llvm::SDVTList, llvm::ArrayRef<llvm::SDValue>) () from
/usr/lib/llvm-6.0/lib/libLLVM-6.0.so.1
...
#17 0x00007ffff1a8426f in llvm::legacy::PassManagerImpl::run(llvm::Module&) ()
from /usr/lib/llvm-6.0/lib/libLLVM-6.0.so.1
#18 0x00007ffff550049d in pocl_llvm_codegen (device=<optimized out>,
modp=0x55555575ead0, output=0x7fffffffbc58, output_size=0x7fffffffbc60)
at
/home/sydney/Documents/HEIG/semestre6/TB/travaux/pocl/lib/CL/pocl_llvm_wg.cc:641
#19 0x00007ffff54984d6 in llvm_codegen (tmpdir=0x5555557c1060
"/home/sydney/.cache/pocl/kcache/IN/BIMGDLHBINIBKJMIBLCOAHLEJNCOOOOBDJFJN/vecadd/4000-1-1",
kernel=0x5555557c0f10, device=0x555555758e20,
local_x=4000, local_y=1, local_z=1) at
/home/sydney/Documents/HEIG/semestre6/TB/travaux/pocl/lib/CL/devices/common.c:131
#20 0x00007ffff54a4114 in pocl_check_dlhandle_cache (cmd=0x55555575c310,
initial_refcount=1) at
/home/sydney/Documents/HEIG/semestre6/TB/travaux/pocl/lib/CL/devices/common.c:961
#21 0x00007ffff54ac763 in pocl_basic_submit (node=0x55555575c310,
cq=0x55555575b750) at
/home/sydney/Documents/HEIG/semestre6/TB/travaux/pocl/lib/CL/devices/basic/basic.c:1136
#22 0x00007ffff5486798 in pocl_command_enqueue (command_queue=0x55555575b750,
node=0x55555575c310) at
/home/sydney/Documents/HEIG/semestre6/TB/travaux/pocl/lib/CL/pocl_util.c:556
#23 0x00007ffff5472baa in POclEnqueueNDRangeKernel
(command_queue=0x55555575b750, kernel=0x5555557c0f10, work_dim=1,
global_work_offset=0x0, global_work_size=0x7fffffffdd90, local_work_size=0x0,
num_events_in_wait_list=0, event_wait_list=0x0, event=0x0) at
/home/sydney/Documents/HEIG/semestre6/TB/travaux/pocl/lib/CL/clEnqueueNDRangeKernel.c:571
#24 0x00005555555556f3 in main () at src/vect_add.c:199
For now, I know that I need to set the two variables llvm_target_triple and
llvm_cpu of a device struct so that LLVM is told to generate a binary targeted
for ARMv8. Currently, it is set to armv8-linux-gnueabihf-cortex-a53. Also, I
made the necessary work to compile the kernel builtin functions with the same
llvm triple. I am currently wondering if my assumptions about these two
variables are wrong, that is, are they both the only variables to set. Am I
forgetting something ?
What I figured out is that the parallel.bc file dumped by POCL is malformed.
Indeed, the llvm tool llvm-dis throws "LLVM error : invalid record (Producer :
LLVM6.0, LLVM 6.0)". So I can't even figure out what's inside this file.
If you need more informations or if I missed something, feel free to let me
know.
Thank you in advance,
Sydney
_______________________________________________
pocl-devel mailing list
[email protected]
https://lists.sourceforge.net/lists/listinfo/pocl-devel
