[MLIR][LLVM] Add distinct identifier to DICompileUnit attribute (#77070)

This commit adds a distinct attribute parameter to the DICompileUnit to
enable the modeling of distinctness. LLVM requires DICompileUnits to be
distinct and there are cases where one gets two equivalent compilation
units but LLVM still requires differentiates them. We observed such
cases for combinations of LTO and inline functions.

This patch also changes the DIScopeForLLVMFuncOp pass to a module pass,
to ensure that only one distinct DICompileUnit is created, instead of
one for each function.

flang/lib/Optimizer/Transforms/AddDebugFoundation.cpp

@@ -65,9 +65,9 @@ void AddDebugFoundationPass::runOnOperation() {
   mlir::LLVM::DIFileAttr fileAttr = getFileAttr(inputFilePath);
   mlir::StringAttr producer = mlir::StringAttr::get(context, "Flang");
   mlir::LLVM::DICompileUnitAttr cuAttr = mlir::LLVM::DICompileUnitAttr::get(
-      context, llvm::dwarf::getLanguage("DW_LANG_Fortran95"), fileAttr,
-      producer, /*isOptimized=*/false,
-      mlir::LLVM::DIEmissionKind::LineTablesOnly);
+      context, mlir::DistinctAttr::create(mlir::UnitAttr::get(context)),
+      llvm::dwarf::getLanguage("DW_LANG_Fortran95"), fileAttr, producer,
+      /*isOptimized=*/false, mlir::LLVM::DIEmissionKind::LineTablesOnly);
 
   module.walk([&](mlir::func::FuncOp funcOp) {
     mlir::Location l = funcOp->getLoc();

flang/test/Transforms/debug-line-table-existing.fir

@@ -12,7 +12,7 @@ module attributes {} {
 #di_file = #llvm.di_file<"simple.f90" in "/home/user01/llvm-project/build_release">
 #loc = loc("/home/user01/llvm-project/build_release/simple.f90":0:0)
 #loc1 = loc("/home/user01/llvm-project/build_release/simple.f90":1:1)
-#di_compile_unit = #llvm.di_compile_unit<sourceLanguage = DW_LANG_Fortran95, file = #di_file, producer = "Flang", isOptimized = false, emissionKind = LineTablesOnly>
+#di_compile_unit = #llvm.di_compile_unit<id = distinct[0]<>, sourceLanguage = DW_LANG_Fortran95, file = #di_file, producer = "Flang", isOptimized = false, emissionKind = LineTablesOnly>
 #di_subroutine_type = #llvm.di_subroutine_type<callingConvention = DW_CC_normal, types = #di_basic_type, #di_basic_type>
 #di_subprogram = #llvm.di_subprogram<compileUnit = #di_compile_unit, scope = #di_file, name = "_QPs1", linkageName = "_QPs1", file = #di_file, line = 1, scopeLine = 1, subprogramFlags = Definition, type = #di_subroutine_type>
 #loc2 = loc(fused<#di_subprogram>[#loc1])

flang/test/Transforms/debug-line-table-inc-file.fir

@@ -30,7 +30,7 @@ module attributes {} {
 // CHECK: #[[MODULE_LOC]] = loc("{{.*}}simple.f90":0:0)
 // CHECK: #[[LOC_INC_FILE:.*]] = loc("{{.*}}inc.f90":1:1)
 // CHECK: #[[LOC_FILE:.*]] = loc("{{.*}}simple.f90":3:1)
-// CHECK: #[[DI_CU:.*]] = #llvm.di_compile_unit<sourceLanguage = DW_LANG_Fortran95, file = #[[DI_FILE]], producer = "Flang", isOptimized = false, emissionKind = LineTablesOnly>
+// CHECK: #[[DI_CU:.*]] = #llvm.di_compile_unit<id = distinct[0]<>, sourceLanguage = DW_LANG_Fortran95, file = #[[DI_FILE]], producer = "Flang", isOptimized = false, emissionKind = LineTablesOnly>
 // CHECK: #[[DI_SP_INC:.*]] = #llvm.di_subprogram<compileUnit = #[[DI_CU]], scope = #[[DI_FILE]], name = "_QPsinc", linkageName = "_QPsinc", file = #[[DI_INC_FILE]], {{.*}}>
 // CHECK: #[[DI_SP:.*]] = #llvm.di_subprogram<compileUnit = #[[DI_CU]], scope = #[[DI_FILE]], name = "_QQmain", linkageName = "_QQmain", file = #[[DI_FILE]], {{.*}}>
 // CHECK: #[[FUSED_LOC_INC_FILE]] = loc(fused<#[[DI_SP_INC]]>[#[[LOC_INC_FILE]]])

flang/test/Transforms/debug-line-table.fir

@@ -18,7 +18,7 @@ module attributes { fir.defaultkind = "a1c4d8i4l4r4", fir.kindmap = "", llvm.dat
 // CHECK: #di_file = #llvm.di_file<"[[FILE_NAME:.*]]" in "[[DIR_NAME:.*]]">
 // CHECK: #[[MODULE_LOC]] = loc("[[DIR_NAME]]/[[FILE_NAME]]":1:1)
 // CHECK: #[[SB_LOC]] = loc("./simple.f90":2:1)
-// CHECK: #di_compile_unit = #llvm.di_compile_unit<sourceLanguage = DW_LANG_Fortran95, file = #di_file, producer = "Flang", isOptimized = false, emissionKind = LineTablesOnly>
+// CHECK: #di_compile_unit = #llvm.di_compile_unit<id = distinct[0]<>, sourceLanguage = DW_LANG_Fortran95, file = #di_file, producer = "Flang", isOptimized = false, emissionKind = LineTablesOnly>
 // CHECK: #di_subroutine_type = #llvm.di_subroutine_type<callingConvention = DW_CC_normal, types = #di_basic_type, #di_basic_type>
 // CHECK: #di_subprogram = #llvm.di_subprogram<compileUnit = #di_compile_unit, scope = #di_file, name = "[[SB_NAME]]", linkageName = "[[SB_NAME]]", file = #di_file, line = 1, scopeLine = 1, subprogramFlags = Definition, type = #di_subroutine_type>
 // CHECK: #[[FUSED_SB_LOC]] = loc(fused<#di_subprogram>[#[[SB_LOC]]])

mlir/examples/toy/Ch6/toyc.cpp

@@ -187,8 +187,7 @@ int loadAndProcessMLIR(mlir::MLIRContext &context,
     // This is necessary to have line tables emitted and basic
     // debugger working. In the future we will add proper debug information
     // emission directly from our frontend.
-    pm.addNestedPass<mlir::LLVM::LLVMFuncOp>(
-        mlir::LLVM::createDIScopeForLLVMFuncOpPass());
+    pm.addPass(mlir::LLVM::createDIScopeForLLVMFuncOpPass());
   }
 
   if (mlir::failed(pm.run(*module)))

mlir/examples/toy/Ch7/toyc.cpp

@@ -188,8 +188,7 @@ int loadAndProcessMLIR(mlir::MLIRContext &context,
     // This is necessary to have line tables emitted and basic
     // debugger working. In the future we will add proper debug information
     // emission directly from our frontend.
-    pm.addNestedPass<mlir::LLVM::LLVMFuncOp>(
-        mlir::LLVM::createDIScopeForLLVMFuncOpPass());
+    pm.addPass(mlir::LLVM::createDIScopeForLLVMFuncOpPass());
   }
 
   if (mlir::failed(pm.run(*module)))

mlir/include/mlir/Dialect/LLVMIR/LLVMAttrDefs.td

@@ -342,6 +342,7 @@ def LLVM_DIBasicTypeAttr : LLVM_Attr<"DIBasicType", "di_basic_type",
 def LLVM_DICompileUnitAttr : LLVM_Attr<"DICompileUnit", "di_compile_unit",
                                        /*traits=*/[], "DIScopeAttr"> {
   let parameters = (ins
+    "DistinctAttr":$id,
     LLVM_DILanguageParameter:$sourceLanguage,
     "DIFileAttr":$file,
     OptionalParameter<"StringAttr">:$producer,

mlir/include/mlir/Dialect/LLVMIR/Transforms/Passes.td

@@ -66,7 +66,7 @@ def NVVMOptimizeForTarget : Pass<"llvm-optimize-for-nvvm-target"> {
   let constructor = "::mlir::NVVM::createOptimizeForTargetPass()";
 }
 
-def DIScopeForLLVMFuncOp : Pass<"ensure-debug-info-scope-on-llvm-func", "LLVM::LLVMFuncOp"> {
+def DIScopeForLLVMFuncOp : Pass<"ensure-debug-info-scope-on-llvm-func", "::mlir::ModuleOp"> {
   let summary = "Materialize LLVM debug info subprogram attribute on every LLVMFuncOp";
   let description = [{
     Having a debug info subprogram attribute on a function is required for

mlir/lib/Dialect/LLVMIR/Transforms/DIScopeForLLVMFuncOp.cpp

@@ -34,69 +34,93 @@ static FileLineColLoc extractFileLoc(Location loc) {
   return FileLineColLoc();
 }
 
+/// Creates a DISubprogramAttr with the provided compile unit and attaches it
+/// to the function. Does nothing when the function already has an attached
+/// subprogram.
+static void addScopeToFunction(LLVM::LLVMFuncOp llvmFunc,
+                               LLVM::DICompileUnitAttr compileUnitAttr) {
+
+  Location loc = llvmFunc.getLoc();
+  if (loc->findInstanceOf<mlir::FusedLocWith<LLVM::DISubprogramAttr>>())
+    return;
+
+  MLIRContext *context = llvmFunc->getContext();
+
+  // Filename, line and colmun to associate to the function.
+  LLVM::DIFileAttr fileAttr;
+  int64_t line = 1, col = 1;
+  FileLineColLoc fileLoc = extractFileLoc(loc);
+  if (!fileLoc && compileUnitAttr) {
+    fileAttr = compileUnitAttr.getFile();
+  } else if (!fileLoc) {
+    fileAttr = LLVM::DIFileAttr::get(context, "<unknown>", "");
+  } else {
+    line = fileLoc.getLine();
+    col = fileLoc.getColumn();
+    StringRef inputFilePath = fileLoc.getFilename().getValue();
+    fileAttr =
+        LLVM::DIFileAttr::get(context, llvm::sys::path::filename(inputFilePath),
+                              llvm::sys::path::parent_path(inputFilePath));
+  }
+  auto subroutineTypeAttr =
+      LLVM::DISubroutineTypeAttr::get(context, llvm::dwarf::DW_CC_normal, {});
+
+  StringAttr funcNameAttr = llvmFunc.getNameAttr();
+  auto subprogramAttr = LLVM::DISubprogramAttr::get(
+      context, compileUnitAttr, fileAttr, funcNameAttr, funcNameAttr, fileAttr,
+      /*line=*/line,
+      /*scopeline=*/col,
+      LLVM::DISubprogramFlags::Definition | LLVM::DISubprogramFlags::Optimized,
+      subroutineTypeAttr);
+  llvmFunc->setLoc(FusedLoc::get(context, {loc}, subprogramAttr));
+}
+
 namespace {
 /// Add a debug info scope to LLVMFuncOp that are missing it.
 struct DIScopeForLLVMFuncOp
     : public LLVM::impl::DIScopeForLLVMFuncOpBase<DIScopeForLLVMFuncOp> {
   void runOnOperation() override {
-    LLVM::LLVMFuncOp llvmFunc = getOperation();
-    Location loc = llvmFunc.getLoc();
-    if (loc->findInstanceOf<mlir::FusedLocWith<LLVM::DISubprogramAttr>>())
-      return;
+    ModuleOp module = getOperation();
+    Location loc = module.getLoc();
 
     MLIRContext *context = &getContext();
 
     // To find a DICompileUnitAttr attached to a parent (the module for
     // example), otherwise create a default one.
+    // Find a DICompileUnitAttr attached to the module, otherwise create a
+    // default one.
     LLVM::DICompileUnitAttr compileUnitAttr;
-    if (ModuleOp module = llvmFunc->getParentOfType<ModuleOp>()) {
-      auto fusedCompileUnitAttr =
-          module->getLoc()
-              ->findInstanceOf<mlir::FusedLocWith<LLVM::DICompileUnitAttr>>();
-      if (fusedCompileUnitAttr)
-        compileUnitAttr = fusedCompileUnitAttr.getMetadata();
-    }
-
-    // Filename, line and colmun to associate to the function.
-    LLVM::DIFileAttr fileAttr;
-    int64_t line = 1, col = 1;
-    FileLineColLoc fileLoc = extractFileLoc(loc);
-    if (!fileLoc && compileUnitAttr) {
-      fileAttr = compileUnitAttr.getFile();
-    } else if (!fileLoc) {
-      fileAttr = LLVM::DIFileAttr::get(context, "<unknown>", "");
+    auto fusedCompileUnitAttr =
+        module->getLoc()
+            ->findInstanceOf<mlir::FusedLocWith<LLVM::DICompileUnitAttr>>();
+    if (fusedCompileUnitAttr) {
+      compileUnitAttr = fusedCompileUnitAttr.getMetadata();
     } else {
-      line = fileLoc.getLine();
-      col = fileLoc.getColumn();
-      StringRef inputFilePath = fileLoc.getFilename().getValue();
-      fileAttr = LLVM::DIFileAttr::get(
-          context, llvm::sys::path::filename(inputFilePath),
-          llvm::sys::path::parent_path(inputFilePath));
-    }
-    if (!compileUnitAttr) {
+      LLVM::DIFileAttr fileAttr;
+      if (FileLineColLoc fileLoc = extractFileLoc(loc)) {
+        StringRef inputFilePath = fileLoc.getFilename().getValue();
+        fileAttr = LLVM::DIFileAttr::get(
+            context, llvm::sys::path::filename(inputFilePath),
+            llvm::sys::path::parent_path(inputFilePath));
+      } else {
+        fileAttr = LLVM::DIFileAttr::get(context, "<unknown>", "");
+      }
+
       compileUnitAttr = LLVM::DICompileUnitAttr::get(
-          context, llvm::dwarf::DW_LANG_C, fileAttr,
-          StringAttr::get(context, "MLIR"), /*isOptimized=*/true,
-          LLVM::DIEmissionKind::LineTablesOnly);
+          context, DistinctAttr::create(UnitAttr::get(context)),
+          llvm::dwarf::DW_LANG_C, fileAttr, StringAttr::get(context, "MLIR"),
+          /*isOptimized=*/true, LLVM::DIEmissionKind::LineTablesOnly);
     }
-    auto subroutineTypeAttr =
-        LLVM::DISubroutineTypeAttr::get(context, llvm::dwarf::DW_CC_normal, {});
-
-    StringAttr funcNameAttr = llvmFunc.getNameAttr();
-    auto subprogramAttr =
-        LLVM::DISubprogramAttr::get(context, compileUnitAttr, fileAttr,
-                                    funcNameAttr, funcNameAttr, fileAttr,
-                                    /*line=*/line,
-                                    /*scopeline=*/col,
-                                    LLVM::DISubprogramFlags::Definition |
-                                        LLVM::DISubprogramFlags::Optimized,
-                                    subroutineTypeAttr);
-    llvmFunc->setLoc(FusedLoc::get(context, {loc}, subprogramAttr));
+
+    // Create subprograms for each function with the same distinct compile unit.
+    module.walk([&](LLVM::LLVMFuncOp func) {
+      addScopeToFunction(func, compileUnitAttr);
+    });
   }
 };
 
 } // end anonymous namespace
 
 std::unique_ptr<Pass> mlir::LLVM::createDIScopeForLLVMFuncOpPass() {
   return std::make_unique<DIScopeForLLVMFuncOp>();
-}
+}

mlir/lib/Target/LLVMIR/DebugImporter.cpp

@@ -50,10 +50,11 @@ DIBasicTypeAttr DebugImporter::translateImpl(llvm::DIBasicType *node) {
 DICompileUnitAttr DebugImporter::translateImpl(llvm::DICompileUnit *node) {
   std::optional<DIEmissionKind> emissionKind =
       symbolizeDIEmissionKind(node->getEmissionKind());
-  return DICompileUnitAttr::get(context, node->getSourceLanguage(),
-                                translate(node->getFile()),
-                                getStringAttrOrNull(node->getRawProducer()),
-                                node->isOptimized(), emissionKind.value());
+  return DICompileUnitAttr::get(
+      context, DistinctAttr::create(UnitAttr::get(context)),
+      node->getSourceLanguage(), translate(node->getFile()),
+      getStringAttrOrNull(node->getRawProducer()), node->isOptimized(),
+      emissionKind.value());
 }
 
 DICompositeTypeAttr DebugImporter::translateImpl(llvm::DICompositeType *node) {

mlir/test/Dialect/LLVMIR/add-debuginfo-func-scope.mlir

@@ -1,18 +1,17 @@
-// RUN: mlir-opt %s --pass-pipeline="builtin.module(llvm.func(ensure-debug-info-scope-on-llvm-func))" --split-input-file --mlir-print-debuginfo | FileCheck %s
-
-
+// RUN: mlir-opt %s --pass-pipeline="builtin.module(ensure-debug-info-scope-on-llvm-func)" --split-input-file --mlir-print-debuginfo | FileCheck %s
 
 // CHECK-LABEL: llvm.func @func_no_debug()
 // CHECK: llvm.return loc(#loc
 // CHECK: loc(#loc[[LOC:[0-9]+]])
 // CHECK: #di_file = #llvm.di_file<"<unknown>" in "">
 // CHECK: #di_subprogram = #llvm.di_subprogram<compileUnit = #di_compile_unit, scope = #di_file, name = "func_no_debug", linkageName = "func_no_debug", file = #di_file, line = 1, scopeLine = 1, subprogramFlags = "Definition|Optimized", type = #di_subroutine_type>
 // CHECK: #loc[[LOC]] = loc(fused<#di_subprogram>
-llvm.func @func_no_debug() {
-  llvm.return loc(unknown)
+module {
+  llvm.func @func_no_debug() {
+    llvm.return loc(unknown)
+  } loc(unknown)
 } loc(unknown)
 
-
 // -----
 
 // Test that existing debug info is not overwritten.
@@ -31,7 +30,7 @@ module {
 #di_subroutine_type = #llvm.di_subroutine_type<callingConvention = DW_CC_normal>
 #loc = loc("foo":0:0)
 #loc1 = loc(unknown)
-#di_compile_unit = #llvm.di_compile_unit<sourceLanguage = DW_LANG_C, file = #di_file, producer = "MLIR", isOptimized = true, emissionKind = LineTablesOnly>
+#di_compile_unit = #llvm.di_compile_unit<id = distinct[0]<>, sourceLanguage = DW_LANG_C, file = #di_file, producer = "MLIR", isOptimized = true, emissionKind = LineTablesOnly>
 #di_subprogram = #llvm.di_subprogram<compileUnit = #di_compile_unit, scope = #di_file, name = "func_with_debug", linkageName = "func_with_debug", file = #di_file, line = 42, scopeLine = 42, subprogramFlags = "Definition|Optimized", type = #di_subroutine_type>
 #loc2 = loc(fused<#di_subprogram>[#loc1])
 
@@ -45,7 +44,7 @@ module {
 // CHECK-DAG: #[[DI_FILE_MODULE:.+]] = #llvm.di_file<"bar.mlir" in "baz">
 // CHECK-DAG: #[[DI_FILE_FUNC:.+]] = #llvm.di_file<"file.mlir" in ""> 
 // CHECK-DAG: #loc[[FUNCFILELOC:[0-9]+]] = loc("file.mlir":9:8)
-// CHECK-DAG: #di_compile_unit = #llvm.di_compile_unit<sourceLanguage = DW_LANG_C, file = #[[DI_FILE_MODULE]], producer = "MLIR", isOptimized = true, emissionKind = LineTablesOnly>
+// CHECK-DAG: #di_compile_unit = #llvm.di_compile_unit<id = distinct[0]<>, sourceLanguage = DW_LANG_C, file = #[[DI_FILE_MODULE]], producer = "MLIR", isOptimized = true, emissionKind = LineTablesOnly>
 // CHECK-DAG: #di_subprogram = #llvm.di_subprogram<compileUnit = #di_compile_unit, scope = #[[DI_FILE_FUNC]], name = "propagate_compile_unit", linkageName = "propagate_compile_unit", file = #[[DI_FILE_FUNC]], line = 9, scopeLine = 8, subprogramFlags = "Definition|Optimized", type = #di_subroutine_type>
 // CHECK-DAG: #loc[[MODULELOC]] = loc(fused<#di_compile_unit>[#loc])
 // CHECK-DAG: #loc[[FUNCLOC]] = loc(fused<#di_subprogram>[#loc[[FUNCFILELOC]]
@@ -55,5 +54,20 @@ module {
   } loc("file.mlir":9:8)
 } loc(#loc)
 #di_file = #llvm.di_file<"bar.mlir" in "baz">
-#di_compile_unit = #llvm.di_compile_unit<sourceLanguage = DW_LANG_C, file = #di_file, producer = "MLIR", isOptimized = true, emissionKind = LineTablesOnly>
+#di_compile_unit = #llvm.di_compile_unit<id = distinct[0]<>, sourceLanguage = DW_LANG_C, file = #di_file, producer = "MLIR", isOptimized = true, emissionKind = LineTablesOnly>
 #loc = loc(fused<#di_compile_unit>["foo.mlir":2:1])
+
+// -----
+
+// Test that only one compile unit is created.
+// CHECK-LABEL: module @multiple_funcs
+// CHECK: llvm.di_compile_unit
+// CHECK-NOT: llvm.di_compile_unit
+module @multiple_funcs {
+  llvm.func @func0() {
+    llvm.return loc(unknown)
+  } loc(unknown)
+  llvm.func @func1() {
+    llvm.return loc(unknown)
+  } loc(unknown)
+} loc(unknown)

mlir/test/Dialect/LLVMIR/call-location.mlir

@@ -2,7 +2,7 @@
 
 #di_file = #llvm.di_file<"file.cpp" in "/folder/">
 #di_compile_unit = #llvm.di_compile_unit<
-  sourceLanguage = DW_LANG_C_plus_plus_14, file = #di_file,
+  id = distinct[0]<>, sourceLanguage = DW_LANG_C_plus_plus_14, file = #di_file,
   isOptimized = true, emissionKind = Full
 >
 #di_subprogram = #llvm.di_subprogram<

mlir/test/Dialect/LLVMIR/debuginfo.mlir

@@ -3,10 +3,10 @@
 // CHECK-DAG: #[[FILE:.*]] = #llvm.di_file<"debuginfo.mlir" in "/test/">
 #file = #llvm.di_file<"debuginfo.mlir" in "/test/">
 
-// CHECK-DAG: #[[CU:.*]] = #llvm.di_compile_unit<sourceLanguage = DW_LANG_C, file = #[[FILE]], producer = "MLIR", isOptimized = true, emissionKind = Full>
+// CHECK-DAG: #[[CU:.*]] = #llvm.di_compile_unit<id = distinct[0]<>, sourceLanguage = DW_LANG_C, file = #[[FILE]], producer = "MLIR", isOptimized = true, emissionKind = Full>
 #cu = #llvm.di_compile_unit<
-  sourceLanguage = DW_LANG_C, file = #file, producer = "MLIR",
-  isOptimized = true, emissionKind = Full
+  id = distinct[0]<>, sourceLanguage = DW_LANG_C, file = #file,
+  producer = "MLIR", isOptimized = true, emissionKind = Full
 >
 
 // CHECK-DAG: #[[NULL:.*]] = #llvm.di_null_type

mlir/test/Dialect/LLVMIR/global.mlir

@@ -263,7 +263,7 @@ llvm.mlir.global @target_fail(0 : i64) : !llvm.target<"spirv.Image", i32, 0>
 
 // CHECK-DAG: #[[TYPE:.*]] = #llvm.di_basic_type<tag = DW_TAG_base_type, name = "uint64_t", sizeInBits = 64, encoding = DW_ATE_unsigned>
 // CHECK-DAG: #[[FILE:.*]] = #llvm.di_file<"not" in "existence">
-// CHECK-DAG: #[[CU:.*]] = #llvm.di_compile_unit<sourceLanguage = DW_LANG_C, file = #[[FILE]], producer = "MLIR", isOptimized = true, emissionKind = Full>
+// CHECK-DAG: #[[CU:.*]] = #llvm.di_compile_unit<id = distinct[0]<>, sourceLanguage = DW_LANG_C, file = #[[FILE]], producer = "MLIR", isOptimized = true, emissionKind = Full>
 // CHECK-DAG: #[[GVAR0:.*]] = #llvm.di_global_variable<scope = #[[CU]], name = "global_with_expr_1", linkageName = "global_with_expr_1", file = #[[FILE]], line = 370, type = #[[TYPE]], isLocalToUnit = true, isDefined = true, alignInBits = 8>
 // CHECK-DAG: #[[GVAR1:.*]] = #llvm.di_global_variable<scope = #[[CU]], name = "global_with_expr_2", linkageName = "global_with_expr_2", file = #[[FILE]], line = 371, type = #[[TYPE]], isLocalToUnit = true, isDefined = true, alignInBits = 8>
 // CHECK-DAG: #[[GVAR2:.*]] = #llvm.di_global_variable<scope = #[[CU]], name = "global_with_expr_3", linkageName = "global_with_expr_3", file = #[[FILE]], line = 372, type = #[[TYPE]], isLocalToUnit = true, isDefined = true, alignInBits = 8>
@@ -278,7 +278,7 @@ llvm.mlir.global @target_fail(0 : i64) : !llvm.target<"spirv.Image", i32, 0>
 // CHECK-DAG:   llvm.mlir.global external @global_with_expr4() {addr_space = 0 : i32, dbg_expr = #[[EXPR3]]} : i64
 
 #di_file = #llvm.di_file<"not" in "existence">
-#di_compile_unit = #llvm.di_compile_unit<sourceLanguage = DW_LANG_C, file = #di_file, producer = "MLIR", isOptimized = true, emissionKind = Full>
+#di_compile_unit = #llvm.di_compile_unit<id = distinct[0]<>, sourceLanguage = DW_LANG_C, file = #di_file, producer = "MLIR", isOptimized = true, emissionKind = Full>
 #di_basic_type = #llvm.di_basic_type<tag = DW_TAG_base_type, name = "uint64_t", sizeInBits = 64, encoding = DW_ATE_unsigned>
 llvm.mlir.global external @global_with_expr1() {addr_space = 0 : i32, dbg_expr = #llvm.di_global_variable_expression<var = <scope = #di_compile_unit, name = "global_with_expr_1", linkageName = "global_with_expr_1", file = #di_file, line = 370, type = #di_basic_type, isLocalToUnit = true, isDefined = true, alignInBits = 8>, expr = <>>} : i64
 llvm.mlir.global external @global_with_expr2() {addr_space = 0 : i32, dbg_expr = #llvm.di_global_variable_expression<var = <scope = #di_compile_unit, name = "global_with_expr_2", linkageName = "global_with_expr_2", file = #di_file, line = 371, type = #di_basic_type, isLocalToUnit = true, isDefined = true, alignInBits = 8>, expr = <[DW_OP_push_object_address, DW_OP_deref]>>} : i64