[PATCH] D133705: [HIP] Fix unbundling archive

[Yaxun Liu via Phabricator via cfe-commits]

yaxunl created this revision.
yaxunl added a reviewer: tra.
Herald added a project: All.
yaxunl requested review of this revision.

HIP is able to unbundle archive of bundled bitcode.
However currently there are two bugs:

1. archives passed  by -l: are not unbundled.
2. archives passed as input files are not unbundled

Archives passed by -l: should not be prefixed with
prefix `lib` and appended with '.a', but still need to be prefixed with
paths in -L options.

Archives passed as input files should not be prefixed
or appended with anything.


https://reviews.llvm.org/D133705

Files:
  clang/lib/CodeGen/CGExpr.cpp
  clang/test/CodeGenCUDA/bool-range.cu


Index: clang/test/CodeGenCUDA/bool-range.cu
===================================================================
--- /dev/null
+++ clang/test/CodeGenCUDA/bool-range.cu
@@ -0,0 +1,15 @@
+// RUN: %clang_cc1 -emit-llvm %s -O3 -o - -fcuda-is-device \
+// RUN:   -triple nvptx64-unknown-unknown | FileCheck %s
+
+#include "Inputs/cuda.h"
+
+// Make sure bool loaded from memory is truncated and
+// range metadata is not emitted.
+
+// CHECK:  %0 = load i8, i8* %x
+// CHECK:  %1 = and i8 %0, 1
+// CHECK:  store i8 %1, i8* %y
+// CHECK-NOT: !range
+__global__ void test1(bool *x, bool *y) {
+  *y = *x != false;
+}
Index: clang/lib/CodeGen/CGExpr.cpp
===================================================================
--- clang/lib/CodeGen/CGExpr.cpp
+++ clang/lib/CodeGen/CGExpr.cpp
@@ -1791,7 +1791,8 @@
   if (EmitScalarRangeCheck(Load, Ty, Loc)) {
     // In order to prevent the optimizer from throwing away the check, don't
     // attach range metadata to the load.
-  } else if (CGM.getCodeGenOpts().OptimizationLevel > 0)
+  } else if (CGM.getCodeGenOpts().OptimizationLevel > 0 &&
+             !(getLangOpts().CUDA && getLangOpts().CUDAIsDevice))
     if (llvm::MDNode *RangeInfo = getRangeForLoadFromType(Ty))
       Load->setMetadata(llvm::LLVMContext::MD_range, RangeInfo);
 


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D133705.458326.patch
Type: text/x-patch
Size: 1286 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20220912/044f13e5/attachment.bin>