HSAFoundation / CLOC

CL Offline Compiler : Compile OpenCL kernels to HSAIL
Other
49 stars 18 forks source link

Error when using "-g -hsail" option on OpenCL kernel #12

Closed owuntu closed 8 years ago

owuntu commented 8 years ago

Hi, I am getting an error message when trying to use cloc.sh to get hsail text from CL kernel with "-g" flag. For example, for this kernel:

__kernel void trap()
{}

using command:

cloc.sh -hsail -g trap.cl

Then I get this message

#Info:  Version:    cloc.sh 0.9.8
#Info:  Input file: /home/user/workspace/CLWorkSpace/trap.cl
#Info:  Brig file:  /home/user/workspace/CLWorkSpace/trap.brig
#Info:  HSAIL file: /home/user/workspace/CLWorkSpace/trap.hsail
#Info:  Run date:   Mon Dec  7 11:54:20 EST 2015
#Info:  LLVM path:  /opt/amd/cloc/bin
#Step:  Compile(clc2)   cl --> bc ...
#Step:  Link(llvm-link) bc --> lnkd.bc ...
#Step:  Optimize(opt)   lnkd.bc --> opt.bc -O2 ...
#Step:  llc arch=hsail  opt.bc --> hsail ...
#Step:  HSAILasm    hsail --> trap.brig -O2 ...
Using libHSAIL options (append from env variable LIBHSAIL_OPTIONS_APPEND): '-o /home/user/workspace/CLWorkSpace/trap.brig /tmp/cloc5129/trap.hsail -g -include-source'
#Step:  HSAILasm    brig --> trap.hsail ...
Using libHSAIL options (append from env variable LIBHSAIL_OPTIONS_APPEND): '-disassemble -o /home/user/workspace/CLWorkSpace/trap.hsail /home/user/workspace/CLWorkSpace/trap.brig -g -include-source'
Invalid section size: must be a multiple of 4
ERROR:  The following command failed with return code 1.
        HSAILasm -disassemble -o /home/user/workspace/CLWorkSpace/trap.hsail /home/user/workspace/CLWorkSpace/trap.brig

No hsail file is generated.

gregrodgers commented 8 years ago

I am trying to cleanup old issues and found this is still a problem. Sorry I was not aware of these github issues. It appears that the -hsail -g option is still giving an error with cloc 1.0.11. I will look into this more.

Greg

gregrodgers commented 8 years ago

This is a bug in the old HSAILasm. I discovered that the new HSAILasm does not process amd extension instructions without the amd prefix so I cannot move to the new HSAILasm unless I move to a new HLC (HSAIL code generator). That is a fairly big project to move to a new HLC in cloc especially with the focus on HSA code object. I tested an empty kernel with -g option for code object and that works. So can we close this as not fixed?