Prevent output flushing on error

[stuartarchibald]

If the library hits some problem during an action, e.g. cannot relocate symbols, it's fairly common for the output to get repeatedly flushed with large amounts of out empty output, technically this isn't a problem, but if preventable it'd make reading the output more convenient. Thanks.

[scott-linder]

Do you have a minimal example? Currently the only output comes from tools which do not expose options to control when output is flushed, but as part of #2 we should replace much of this output to standard streams with LOG objects.

[stuartarchibald]

Unfortunately not. I have a tower of half working things whilst switching out the Numba ROC target compilation pipeline to use comgr. The place I observe this is when I do the following:

• Create a dataset with a single LLVM IR source and 9 bitcode sources:

  • LLVM IR is generated by Numba

  • Bitcodes are from the ROCm 1.9.x release:

    • "opencl.amdgcn.bc",
    • "ocml.amdgcn.bc",
    • "ockl.amdgcn.bc",
    • "oclc_correctly_rounded_sqrt_off.amdgcn.bc",
    • "oclc_daz_opt_off.amdgcn.bc",
    • "oclc_finite_only_off.amdgcn.bc",
    • "oclc_isa_version_803.amdgcn.bc",
    • "oclc_unsafe_math_off.amdgcn.bc",
    • "irif.amdgcn.bc

• Perform an action kind of AMD_COMGR_ACTION_ASSEMBLE_SOURCE_TO_RELOCATABLE on the dataset

• Perform an action kind of AMD_COMGR_ACTION_LINK_RELOCATABLE_TO_RELOCATABLE on the resulting dataset

• Perform an action kind of AMD_COMGR_ACTION_LINK_RELOCATABLE_TO_EXECUTABLE on the resulting dataset. At which point about 5 screens of blank flushing takes place. Scrolling up, there's loads of errors like:

  : error: relocation R_AMDGPU_REL32_LO cannot be used against symbol get_local_id(unsigned int); recompile with -fPIC
  >>> defined in /tmp/comgr-7102b3/input/numba_generated_ir.ll.o
  >>> referenced by /tmp/comgr-7102b3/input/numba_generated_ir.ll.o:(hsapy_devfn_numba_2E_roc_2E_tests_2E_hsapy_2E_test_5F_scan_2E_shuffle_5F_up_24_2_2E_int32_2E_int64)
  

  I'm guessing the error is because of the baked in -r in the lld binding there's no way to -shared. Also can't find a way to add -fPIC (which was a bit what this was about https://github.com/RadeonOpenCompute/ROCm-CompilerSupport/issues/5, in that I'm trying to relate the comgr behaviours and source back to a familiar implementation). I can open the above link problem as a new issue if it's best resolved there.

Thanks for your help.

[lamb-j]

I'm guessing this is no longer an issue. I haven't seen or heard of this error recently. Follow up if you're still able to recreate it though!