diff --git a/src/target/llvm/codegen_hexagon.cc b/src/target/llvm/codegen_hexagon.cc index 2f91807b6933..d9d0d1f3d6a4 100644 --- a/src/target/llvm/codegen_hexagon.cc +++ b/src/target/llvm/codegen_hexagon.cc @@ -671,11 +671,9 @@ runtime::Module BuildHexagon(IRModule mod, Target target) { } return vec; }; - std::string llvm_options_str; - if (const Optional llvm_options = target->GetAttr("llvm-options")) { - llvm_options_str = "llvm," + llvm_options.value(); - } else { - llvm_options_str = "llvm"; + std::string llvm_options_str = "llvm"; + if (const auto& llvm_options = target->GetAttr>("llvm-options")) { + for (const String& s : llvm_options.value()) llvm_options_str += "," + s; } // Postprocess the LLVM options string: replace '@' with '=', and ',' with ' '. for (int i = 0, e = llvm_options_str.size(); i != e; ++i) { diff --git a/tests/python/unittest/test_target_codegen_hexagon.py b/tests/python/unittest/test_target_codegen_hexagon.py index 28901b35e75b..c8b48993967b 100644 --- a/tests/python/unittest/test_target_codegen_hexagon.py +++ b/tests/python/unittest/test_target_codegen_hexagon.py @@ -109,6 +109,18 @@ def test_alloc_vtcm(): assert "HexagonBackendFreeVTCM" in calls +def test_llvm_options(): + if not check_prereq_and_setup(): + return + target = tvm.target.hexagon("v66", llvm_options="-hexagon-noopt") + Zero = tvm.te.compute((10,), lambda _: tvm.tir.const(0, "int32")) + s = tvm.te.create_schedule(Zero.op) + tvm.build(s, [Zero], target=target, name="zero") + # Check that BuildHexagon hasn't crashed because of target attribute + # type mismatch. + assert re.search("-hexagon-noopt", str(target)) + + def test_linked_params_codegen(): if not check_prereq_and_setup(): return @@ -176,4 +188,5 @@ def test_linked_params_codegen(): test_basic() test_llvm_target_features() test_alloc_vtcm() + test_llvm_options() test_linked_params_codegen()