; LLVM IR from AdaptiveCpp/AdaptiveCpp :: src/runtime/ocl (ocl_hardware_manager)

; Unoptimized IR (clang -disable-llvm-passes):
target datalayout = "e-m:e-p270:32:32-p271:32:32-p272:64:64-i64:64-i128:128-f80:128-n8:16:32:64-S128"
target triple = "x86_64-unknown-linux-gnu"

%"class.std::unique_ptr" = type { %"struct.std::__uniq_ptr_data" }
%"struct.std::__uniq_ptr_data" = type { %"class.std::__uniq_ptr_impl" }
%"class.std::__uniq_ptr_impl" = type { %"class.std::tuple" }
%"class.std::tuple" = type { %"struct.std::_Tuple_impl" }
%"struct.std::_Tuple_impl" = type { %"struct.std::_Head_base.5" }
%"struct.std::_Head_base.5" = type { ptr }
%"class.std::__cxx11::basic_string" = type { %"struct.std::__cxx11::basic_string<char>::_Alloc_hider", i64, %union.anon }
%"struct.std::__cxx11::basic_string<char>::_Alloc_hider" = type { ptr }
%union.anon = type { i64, [8 x i8] }
%"class.hipsycl::rt::device_id" = type { %"struct.hipsycl::rt::backend_descriptor", i32 }
%"struct.hipsycl::rt::backend_descriptor" = type { i32, i32, i32 }
%"class.hipsycl::rt::ocl_allocator" = type { %"class.hipsycl::rt::backend_allocator", ptr, %"class.hipsycl::rt::device_id" }
%"class.hipsycl::rt::backend_allocator" = type { ptr }
%"class.hipsycl::rt::ocl_hardware_context" = type <{ %"class.hipsycl::rt::hardware_context", i32, i32, %"class.cl::Context", %"class.cl::Device", %"class.std::shared_ptr", %"class.hipsycl::rt::ocl_allocator", i8, [7 x i8] }>
%"class.hipsycl::rt::hardware_context" = type { ptr }
%"class.cl::Context" = type { %"class.cl::detail::Wrapper" }
%"class.cl::detail::Wrapper" = type { ptr }
%"class.cl::Device" = type { %"class.cl::detail::Wrapper.base", [7 x i8] }
%"class.cl::detail::Wrapper.base" = type <{ ptr, i8 }>
%"class.std::shared_ptr" = type { %"class.std::__shared_ptr" }
%"class.std::__shared_ptr" = type { ptr, %"class.std::__shared_count" }
%"class.std::__shared_count" = type { ptr }

@.str.7 = external hidden unnamed_addr constant [33 x i8], align 1
@.str.8 = external hidden unnamed_addr constant [15 x i8], align 1
@.str.9 = external hidden unnamed_addr constant [128 x i8], align 1
@.str.10 = external hidden unnamed_addr constant [89 x i8], align 1

declare i32 @__gxx_personality_v0(...)

; Function Attrs: nocallback nofree nosync nounwind willreturn memory(argmem: readwrite)
declare void @llvm.lifetime.start.p0(i64 immarg, ptr nocapture) #0

; Function Attrs: mustprogress nounwind uwtable
declare void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32)) unnamed_addr #1 align 2

; Function Attrs: nocallback nofree nosync nounwind willreturn memory(argmem: readwrite)
declare void @llvm.lifetime.end.p0(i64 immarg, ptr nocapture) #0

; Function Attrs: mustprogress nounwind uwtable
declare dso_local void @_ZN7hipsycl2rt17backend_allocatorD2Ev(ptr noundef nonnull align 8 dereferenceable(8)) unnamed_addr #1 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef ptr @_ZNKSt19__shared_ptr_accessIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2ELb0ELb0EEptEv(ptr noundef nonnull align 1 dereferenceable(1)) #1 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef ptr @_ZNKSt12__shared_ptrIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2EE3getEv(ptr noundef nonnull align 8 dereferenceable(16)) #1 align 2

; Function Attrs: mustprogress uwtable
define dso_local void @_ZN7hipsycl2rt20ocl_hardware_context14init_allocatorEPNS0_20ocl_hardware_managerE(ptr noundef nonnull align 8 dereferenceable(89) %0, ptr noundef %1) #2 align 2 personality ptr @__gxx_personality_v0 {
  %3 = alloca ptr, align 8
  %4 = alloca ptr, align 8
  %5 = alloca %"class.std::unique_ptr", align 8
  %6 = alloca ptr, align 8
  %7 = alloca i32, align 4
  %8 = alloca %"class.std::unique_ptr", align 8
  %9 = alloca %"class.std::__cxx11::basic_string", align 8
  %10 = alloca %"class.std::__cxx11::basic_string", align 8
  %11 = alloca %"class.hipsycl::rt::device_id", align 4
  %12 = alloca %"struct.hipsycl::rt::backend_descriptor", align 4
  %13 = alloca { i64, i32 }, align 4
  %14 = alloca %"class.hipsycl::rt::ocl_allocator", align 8
  %15 = alloca %"class.hipsycl::rt::device_id", align 4
  store ptr %0, ptr %3, align 8, !tbaa !5
  store ptr %1, ptr %4, align 8, !tbaa !5
  %16 = load ptr, ptr %3, align 8
  call void @llvm.lifetime.start.p0(i64 8, ptr %5) #7
  %17 = load ptr, ptr %4, align 8, !tbaa !5
  %18 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 1
  %19 = load i32, ptr %18, align 8, !tbaa !9
  call void @_ZN7hipsycl2rt7ocl_usm20from_intel_extensionEPNS0_20ocl_hardware_managerEi(ptr dead_on_unwind writable sret(%"class.std::unique_ptr") align 8 %5, ptr noundef %17, i32 noundef %19)
  %20 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 5
  %21 = invoke noundef nonnull align 8 dereferenceable(16) ptr @_ZNSt10shared_ptrIN7hipsycl2rt7ocl_usmEEaSIS2_St14default_deleteIS2_EEENSt9enable_ifIXsr13is_assignableIRSt12__shared_ptrIS2_LN9__gnu_cxx12_Lock_policyE2EESt10unique_ptrIT_T0_EEE5valueERS3_E4typeEOSG_(ptr noundef nonnull align 8 dereferenceable(16) %20, ptr noundef nonnull align 8 dereferenceable(8) %5)
          to label %22 unwind label %47

22:                                               ; preds = %2
  call void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8) %5) #7
  call void @llvm.lifetime.end.p0(i64 8, ptr %5) #7
  %23 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 5
  %24 = call noundef ptr @_ZNKSt19__shared_ptr_accessIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2ELb0ELb0EEptEv(ptr noundef nonnull align 1 dereferenceable(1) %23) #7
  %25 = load ptr, ptr %24, align 8, !tbaa !28
  %26 = getelementptr inbounds ptr, ptr %25, i64 2
  %27 = load ptr, ptr %26, align 8
  %28 = call noundef zeroext i1 %27(ptr noundef nonnull align 8 dereferenceable(8) %24)
  br i1 %28, label %75, label %29

29:                                               ; preds = %22
  call void @llvm.lifetime.start.p0(i64 8, ptr %8) #7
  %30 = load ptr, ptr %4, align 8, !tbaa !5
  %31 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 1
  %32 = load i32, ptr %31, align 8, !tbaa !9
  call void @_ZN7hipsycl2rt7ocl_usm28from_fine_grained_system_svmEPNS0_20ocl_hardware_managerEi(ptr dead_on_unwind writable sret(%"class.std::unique_ptr") align 8 %8, ptr noundef %30, i32 noundef %32)
  %33 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 5
  %34 = invoke noundef nonnull align 8 dereferenceable(16) ptr @_ZNSt10shared_ptrIN7hipsycl2rt7ocl_usmEEaSIS2_St14default_deleteIS2_EEENSt9enable_ifIXsr13is_assignableIRSt12__shared_ptrIS2_LN9__gnu_cxx12_Lock_policyE2EESt10unique_ptrIT_T0_EEE5valueERS3_E4typeEOSG_(ptr noundef nonnull align 8 dereferenceable(16) %33, ptr noundef nonnull align 8 dereferenceable(8) %8)
          to label %35 unwind label %51

35:                                               ; preds = %29
  call void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8) %8) #7
  call void @llvm.lifetime.end.p0(i64 8, ptr %8) #7
  %36 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 5
  %37 = call noundef ptr @_ZNKSt19__shared_ptr_accessIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2ELb0ELb0EEptEv(ptr noundef nonnull align 1 dereferenceable(1) %36) #7
  %38 = load ptr, ptr %37, align 8, !tbaa !28
  %39 = getelementptr inbounds ptr, ptr %38, i64 2
  %40 = load ptr, ptr %39, align 8
  %41 = call noundef zeroext i1 %40(ptr noundef nonnull align 8 dereferenceable(8) %37)
  br i1 %41, label %42, label %74

42:                                               ; preds = %35
  %43 = call noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv()
  %44 = call noundef i32 @_ZNK7hipsycl6common13output_stream15get_debug_levelEv(ptr noundef nonnull align 8 dereferenceable(16) %43)
  %45 = icmp sgt i32 2, %44
  br i1 %45, label %46, label %55

46:                                               ; preds = %42
  br label %73

47:                                               ; preds = %2
  %48 = landingpad { ptr, i32 }
          cleanup
  %49 = extractvalue { ptr, i32 } %48, 0
  store ptr %49, ptr %6, align 8
  %50 = extractvalue { ptr, i32 } %48, 1
  store i32 %50, ptr %7, align 4
  call void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8) %5) #7
  call void @llvm.lifetime.end.p0(i64 8, ptr %5) #7
  br label %121

51:                                               ; preds = %29
  %52 = landingpad { ptr, i32 }
          cleanup
  %53 = extractvalue { ptr, i32 } %52, 0
  store ptr %53, ptr %6, align 8
  %54 = extractvalue { ptr, i32 } %52, 1
  store i32 %54, ptr %7, align 4
  call void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8) %8) #7
  call void @llvm.lifetime.end.p0(i64 8, ptr %8) #7
  br label %121

55:                                               ; preds = %42
  %56 = call noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv()
  %57 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZNK7hipsycl6common13output_stream10get_streamEv(ptr noundef nonnull align 8 dereferenceable(16) %56)
  %58 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %57, ptr noundef @.str.7)
  %59 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %58, ptr noundef @.str.8)
  call void @llvm.lifetime.start.p0(i64 32, ptr %9) #7
  %60 = load ptr, ptr %16, align 8, !tbaa !28
  %61 = getelementptr inbounds ptr, ptr %60, i64 4
  %62 = load ptr, ptr %61, align 8
  call void %62(ptr dead_on_unwind writable sret(%"class.std::__cxx11::basic_string") align 8 %9, ptr noundef nonnull align 8 dereferenceable(89) %16)
  %63 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsIcSt11char_traitsIcESaIcEERSt13basic_ostreamIT_T0_ES7_RKNSt7__cxx1112basic_stringIS4_S5_T1_EE(ptr noundef nonnull align 8 dereferenceable(8) %59, ptr noundef nonnull align 8 dereferenceable(32) %9)
          to label %64 unwind label %69

64:                                               ; preds = %55
  %65 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %63, ptr noundef @.str.9)
          to label %66 unwind label %69

66:                                               ; preds = %64
  %67 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZNSolsEPFRSoS_E(ptr noundef nonnull align 8 dereferenceable(8) %65, ptr noundef @_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_)
          to label %68 unwind label %69

68:                                               ; preds = %66
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %9) #7
  call void @llvm.lifetime.end.p0(i64 32, ptr %9) #7
  br label %73

69:                                               ; preds = %66, %64, %55
  %70 = landingpad { ptr, i32 }
          cleanup
  %71 = extractvalue { ptr, i32 } %70, 0
  store ptr %71, ptr %6, align 8
  %72 = extractvalue { ptr, i32 } %70, 1
  store i32 %72, ptr %7, align 4
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %9) #7
  call void @llvm.lifetime.end.p0(i64 32, ptr %9) #7
  br label %121

73:                                               ; preds = %68, %46
  br label %74

74:                                               ; preds = %73, %35
  br label %75

75:                                               ; preds = %74, %22
  %76 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 5
  %77 = call noundef ptr @_ZNKSt19__shared_ptr_accessIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2ELb0ELb0EEptEv(ptr noundef nonnull align 1 dereferenceable(1) %76) #7
  %78 = load ptr, ptr %77, align 8, !tbaa !28
  %79 = getelementptr inbounds ptr, ptr %78, i64 2
  %80 = load ptr, ptr %79, align 8
  %81 = call noundef zeroext i1 %80(ptr noundef nonnull align 8 dereferenceable(8) %77)
  br i1 %81, label %106, label %82

82:                                               ; preds = %75
  %83 = call noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv()
  %84 = call noundef i32 @_ZNK7hipsycl6common13output_stream15get_debug_levelEv(ptr noundef nonnull align 8 dereferenceable(16) %83)
  %85 = icmp sgt i32 2, %84
  br i1 %85, label %86, label %87

86:                                               ; preds = %82
  br label %105

87:                                               ; preds = %82
  %88 = call noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv()
  %89 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZNK7hipsycl6common13output_stream10get_streamEv(ptr noundef nonnull align 8 dereferenceable(16) %88)
  %90 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %89, ptr noundef @.str.7)
  %91 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %90, ptr noundef @.str.8)
  call void @llvm.lifetime.start.p0(i64 32, ptr %10) #7
  %92 = load ptr, ptr %16, align 8, !tbaa !28
  %93 = getelementptr inbounds ptr, ptr %92, i64 4
  %94 = load ptr, ptr %93, align 8
  call void %94(ptr dead_on_unwind writable sret(%"class.std::__cxx11::basic_string") align 8 %10, ptr noundef nonnull align 8 dereferenceable(89) %16)
  %95 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsIcSt11char_traitsIcESaIcEERSt13basic_ostreamIT_T0_ES7_RKNSt7__cxx1112basic_stringIS4_S5_T1_EE(ptr noundef nonnull align 8 dereferenceable(8) %91, ptr noundef nonnull align 8 dereferenceable(32) %10)
          to label %96 unwind label %101

96:                                               ; preds = %87
  %97 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %95, ptr noundef @.str.10)
          to label %98 unwind label %101

98:                                               ; preds = %96
  %99 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZNSolsEPFRSoS_E(ptr noundef nonnull align 8 dereferenceable(8) %97, ptr noundef @_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_)
          to label %100 unwind label %101

100:                                              ; preds = %98
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %10) #7
  call void @llvm.lifetime.end.p0(i64 32, ptr %10) #7
  br label %105

101:                                              ; preds = %98, %96, %87
  %102 = landingpad { ptr, i32 }
          cleanup
  %103 = extractvalue { ptr, i32 } %102, 0
  store ptr %103, ptr %6, align 8
  %104 = extractvalue { ptr, i32 } %102, 1
  store i32 %104, ptr %7, align 4
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %10) #7
  call void @llvm.lifetime.end.p0(i64 32, ptr %10) #7
  br label %121

105:                                              ; preds = %100, %86
  br label %106

106:                                              ; preds = %105, %75
  call void @llvm.lifetime.start.p0(i64 16, ptr %11) #7
  call void @_ZN7hipsycl2rt18backend_descriptorC2ENS0_17hardware_platformENS0_12api_platformE(ptr noundef nonnull align 4 dereferenceable(12) %12, i32 noundef 3, i32 noundef 3)
  %107 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 1
  %108 = load i32, ptr %107, align 8, !tbaa !9
  call void @llvm.memcpy.p0.p0.i64(ptr align 4 %13, ptr align 4 %12, i64 12, i1 false)
  %109 = getelementptr inbounds { i64, i32 }, ptr %13, i32 0, i32 0
  %110 = load i64, ptr %109, align 4
  %111 = getelementptr inbounds { i64, i32 }, ptr %13, i32 0, i32 1
  %112 = load i32, ptr %111, align 4
  call void @_ZN7hipsycl2rt9device_idC1ENS0_18backend_descriptorEi(ptr noundef nonnull align 4 dereferenceable(16) %11, i64 %110, i32 %112, i32 noundef %108)
  call void @llvm.lifetime.start.p0(i64 32, ptr %14) #7
  call void @llvm.memcpy.p0.p0.i64(ptr align 4 %15, ptr align 4 %11, i64 16, i1 false), !tbaa.struct !30
  %113 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 5
  %114 = call noundef ptr @_ZNKSt12__shared_ptrIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2EE3getEv(ptr noundef nonnull align 8 dereferenceable(16) %113) #7
  %115 = getelementptr inbounds { i64, i64 }, ptr %15, i32 0, i32 0
  %116 = load i64, ptr %115, align 4
  %117 = getelementptr inbounds { i64, i64 }, ptr %15, i32 0, i32 1
  %118 = load i64, ptr %117, align 4
  call void @_ZN7hipsycl2rt13ocl_allocatorC1ENS0_9device_idEPNS0_7ocl_usmE(ptr noundef nonnull align 8 dereferenceable(32) %14, i64 %116, i64 %118, ptr noundef %114)
  %119 = getelementptr inbounds %"class.hipsycl::rt::ocl_hardware_context", ptr %16, i32 0, i32 6
  %120 = call noundef nonnull align 8 dereferenceable(32) ptr @_ZN7hipsycl2rt13ocl_allocatoraSEOS1_(ptr noundef nonnull align 8 dereferenceable(32) %119, ptr noundef nonnull align 8 dereferenceable(32) %14) #7
  call void @_ZN7hipsycl2rt17backend_allocatorD2Ev(ptr noundef nonnull align 8 dereferenceable(32) %14) #7
  call void @llvm.lifetime.end.p0(i64 32, ptr %14) #7
  call void @llvm.lifetime.end.p0(i64 16, ptr %11) #7
  ret void

121:                                              ; preds = %101, %69, %51, %47
  %122 = load ptr, ptr %6, align 8
  %123 = load i32, ptr %7, align 4
  %124 = insertvalue { ptr, i32 } poison, ptr %122, 0
  %125 = insertvalue { ptr, i32 } %124, i32 %123, 1
  resume { ptr, i32 } %125
}

declare void @_ZN7hipsycl2rt7ocl_usm20from_intel_extensionEPNS0_20ocl_hardware_managerEi(ptr dead_on_unwind writable sret(%"class.std::unique_ptr") align 8, ptr noundef, i32 noundef) #3

; Function Attrs: mustprogress uwtable
declare dso_local noundef nonnull align 8 dereferenceable(16) ptr @_ZNSt10shared_ptrIN7hipsycl2rt7ocl_usmEEaSIS2_St14default_deleteIS2_EEENSt9enable_ifIXsr13is_assignableIRSt12__shared_ptrIS2_LN9__gnu_cxx12_Lock_policyE2EESt10unique_ptrIT_T0_EEE5valueERS3_E4typeEOSG_(ptr noundef nonnull align 8 dereferenceable(16), ptr noundef nonnull align 8 dereferenceable(8)) #2 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8)) unnamed_addr #1 align 2

declare void @_ZN7hipsycl2rt7ocl_usm28from_fine_grained_system_svmEPNS0_20ocl_hardware_managerEi(ptr dead_on_unwind writable sret(%"class.std::unique_ptr") align 8, ptr noundef, i32 noundef) #3

; Function Attrs: mustprogress uwtable
declare dso_local noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv() #2 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef i32 @_ZNK7hipsycl6common13output_stream15get_debug_levelEv(ptr noundef nonnull align 8 dereferenceable(16)) #1 align 2

; Function Attrs: inlinehint mustprogress uwtable
declare noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8), ptr noundef) #4

; Function Attrs: inlinehint mustprogress uwtable
declare noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsIcSt11char_traitsIcESaIcEERSt13basic_ostreamIT_T0_ES7_RKNSt7__cxx1112basic_stringIS4_S5_T1_EE(ptr noundef nonnull align 8 dereferenceable(8), ptr noundef nonnull align 8 dereferenceable(32)) #4

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef nonnull align 8 dereferenceable(8) ptr @_ZNK7hipsycl6common13output_stream10get_streamEv(ptr noundef nonnull align 8 dereferenceable(16)) #1 align 2

; Function Attrs: mustprogress uwtable
declare noundef nonnull align 8 dereferenceable(8) ptr @_ZNSolsEPFRSoS_E(ptr noundef nonnull align 8 dereferenceable(8), ptr noundef) #2 align 2

; Function Attrs: inlinehint mustprogress uwtable
declare noundef nonnull align 8 dereferenceable(8) ptr @_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_(ptr noundef nonnull align 8 dereferenceable(8)) #4

; Function Attrs: mustprogress nounwind uwtable
declare dso_local void @_ZN7hipsycl2rt18backend_descriptorC2ENS0_17hardware_platformENS0_12api_platformE(ptr noundef nonnull align 4 dereferenceable(12), i32 noundef, i32 noundef) unnamed_addr #1 align 2

declare void @_ZN7hipsycl2rt9device_idC1ENS0_18backend_descriptorEi(ptr noundef nonnull align 4 dereferenceable(16), i64, i32, i32 noundef) unnamed_addr #3

; Function Attrs: nocallback nofree nounwind willreturn memory(argmem: readwrite)
declare void @llvm.memcpy.p0.p0.i64(ptr noalias nocapture writeonly, ptr noalias nocapture readonly, i64, i1 immarg) #5

declare void @_ZN7hipsycl2rt13ocl_allocatorC1ENS0_9device_idEPNS0_7ocl_usmE(ptr noundef nonnull align 8 dereferenceable(32), i64, i64, ptr noundef) unnamed_addr #3

; Function Attrs: inlinehint mustprogress nounwind uwtable
declare dso_local noundef nonnull align 8 dereferenceable(32) ptr @_ZN7hipsycl2rt13ocl_allocatoraSEOS1_(ptr noundef nonnull align 8 dereferenceable(32), ptr noundef nonnull align 8 dereferenceable(32)) #6 align 2

attributes #0 = { nocallback nofree nosync nounwind willreturn memory(argmem: readwrite) }
attributes #1 = { mustprogress nounwind uwtable "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #2 = { mustprogress uwtable "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #3 = { "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #4 = { inlinehint mustprogress uwtable "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #5 = { nocallback nofree nounwind willreturn memory(argmem: readwrite) }
attributes #6 = { inlinehint mustprogress nounwind uwtable "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #7 = { nounwind }

!llvm.module.flags = !{!0, !1, !2, !3}
!llvm.ident = !{!4}

!0 = !{i32 1, !"wchar_size", i32 4}
!1 = !{i32 8, !"PIC Level", i32 2}
!2 = !{i32 7, !"PIE Level", i32 2}
!3 = !{i32 7, !"uwtable", i32 2}
!4 = !{!"clang version 19.1.0 (git@github.com:llvm/llvm-project.git a4bf6cd7cfb1a1421ba92bca9d017b49936c55e4)"}
!5 = !{!6, !6, i64 0}
!6 = !{!"any pointer", !7, i64 0}
!7 = !{!"omnipotent char", !8, i64 0}
!8 = !{!"Simple C++ TBAA"}
!9 = !{!10, !12, i64 8}
!10 = !{!"_ZTSN7hipsycl2rt20ocl_hardware_contextE", !11, i64 0, !12, i64 8, !12, i64 12, !13, i64 16, !15, i64 24, !18, i64 40, !21, i64 56, !17, i64 88}
!11 = !{!"_ZTSN7hipsycl2rt16hardware_contextE"}
!12 = !{!"int", !7, i64 0}
!13 = !{!"_ZTSN2cl7ContextE", !14, i64 0}
!14 = !{!"_ZTSN2cl6detail7WrapperIP11_cl_contextEE", !6, i64 0}
!15 = !{!"_ZTSN2cl6DeviceE", !16, i64 0}
!16 = !{!"_ZTSN2cl6detail7WrapperIP13_cl_device_idEE", !6, i64 0, !17, i64 8}
!17 = !{!"bool", !7, i64 0}
!18 = !{!"_ZTSSt10shared_ptrIN7hipsycl2rt7ocl_usmEE", !19, i64 0}
!19 = !{!"_ZTSSt12__shared_ptrIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2EE", !6, i64 0, !20, i64 8}
!20 = !{!"_ZTSSt14__shared_countILN9__gnu_cxx12_Lock_policyE2EE", !6, i64 0}
!21 = !{!"_ZTSN7hipsycl2rt13ocl_allocatorE", !22, i64 0, !6, i64 8, !23, i64 16}
!22 = !{!"_ZTSN7hipsycl2rt17backend_allocatorE"}
!23 = !{!"_ZTSN7hipsycl2rt9device_idE", !24, i64 0, !12, i64 12}
!24 = !{!"_ZTSN7hipsycl2rt18backend_descriptorE", !25, i64 0, !26, i64 4, !27, i64 8}
!25 = !{!"_ZTSN7hipsycl2rt10backend_idE", !7, i64 0}
!26 = !{!"_ZTSN7hipsycl2rt17hardware_platformE", !7, i64 0}
!27 = !{!"_ZTSN7hipsycl2rt12api_platformE", !7, i64 0}
!28 = !{!29, !29, i64 0}
!29 = !{!"vtable pointer", !8, i64 0}
!30 = !{i64 0, i64 4, !31, i64 4, i64 4, !32, i64 8, i64 4, !33, i64 12, i64 4, !34}
!31 = !{!25, !25, i64 0}
!32 = !{!26, !26, i64 0}
!33 = !{!27, !27, i64 0}
!34 = !{!12, !12, i64 0}


; Optimized IR (opt -passes='default<O3>'):
target datalayout = "e-m:e-p270:32:32-p271:32:32-p272:64:64-i64:64-i128:128-f80:128-n8:16:32:64-S128"
target triple = "x86_64-unknown-linux-gnu"

%"class.std::unique_ptr" = type { %"struct.std::__uniq_ptr_data" }
%"struct.std::__uniq_ptr_data" = type { %"class.std::__uniq_ptr_impl" }
%"class.std::__uniq_ptr_impl" = type { %"class.std::tuple" }
%"class.std::tuple" = type { %"struct.std::_Tuple_impl" }
%"struct.std::_Tuple_impl" = type { %"struct.std::_Head_base.5" }
%"struct.std::_Head_base.5" = type { ptr }
%"class.std::__cxx11::basic_string" = type { %"struct.std::__cxx11::basic_string<char>::_Alloc_hider", i64, %union.anon }
%"struct.std::__cxx11::basic_string<char>::_Alloc_hider" = type { ptr }
%union.anon = type { i64, [8 x i8] }
%"class.hipsycl::rt::device_id" = type { %"struct.hipsycl::rt::backend_descriptor", i32 }
%"struct.hipsycl::rt::backend_descriptor" = type { i32, i32, i32 }
%"class.hipsycl::rt::ocl_allocator" = type { %"class.hipsycl::rt::backend_allocator", ptr, %"class.hipsycl::rt::device_id" }
%"class.hipsycl::rt::backend_allocator" = type { ptr }

@.str.7 = external hidden unnamed_addr constant [33 x i8], align 1
@.str.8 = external hidden unnamed_addr constant [15 x i8], align 1
@.str.9 = external hidden unnamed_addr constant [128 x i8], align 1
@.str.10 = external hidden unnamed_addr constant [89 x i8], align 1

declare i32 @__gxx_personality_v0(...)

; Function Attrs: mustprogress nocallback nofree nosync nounwind willreturn memory(argmem: readwrite)
declare void @llvm.lifetime.start.p0(i64 immarg, ptr nocapture) #0

; Function Attrs: mustprogress nounwind uwtable
declare void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32)) unnamed_addr #1 align 2

; Function Attrs: mustprogress nocallback nofree nosync nounwind willreturn memory(argmem: readwrite)
declare void @llvm.lifetime.end.p0(i64 immarg, ptr nocapture) #0

; Function Attrs: mustprogress nounwind uwtable
declare dso_local void @_ZN7hipsycl2rt17backend_allocatorD2Ev(ptr noundef nonnull align 8 dereferenceable(8)) unnamed_addr #1 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef ptr @_ZNKSt19__shared_ptr_accessIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2ELb0ELb0EEptEv(ptr noundef nonnull align 1 dereferenceable(1)) local_unnamed_addr #1 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef ptr @_ZNKSt12__shared_ptrIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2EE3getEv(ptr noundef nonnull align 8 dereferenceable(16)) local_unnamed_addr #1 align 2

; Function Attrs: mustprogress uwtable
define dso_local void @_ZN7hipsycl2rt20ocl_hardware_context14init_allocatorEPNS0_20ocl_hardware_managerE(ptr noundef nonnull align 8 dereferenceable(89) %0, ptr noundef %1) local_unnamed_addr #2 align 2 personality ptr @__gxx_personality_v0 {
  %3 = alloca %"class.std::unique_ptr", align 8
  %4 = alloca %"class.std::unique_ptr", align 8
  %5 = alloca %"class.std::__cxx11::basic_string", align 8
  %6 = alloca %"class.std::__cxx11::basic_string", align 8
  %7 = alloca %"class.hipsycl::rt::device_id", align 8
  %8 = alloca %"struct.hipsycl::rt::backend_descriptor", align 8
  %9 = alloca %"class.hipsycl::rt::ocl_allocator", align 8
  call void @llvm.lifetime.start.p0(i64 8, ptr nonnull %3) #6
  %10 = getelementptr inbounds i8, ptr %0, i64 8
  %11 = load i32, ptr %10, align 8, !tbaa !5
  call void @_ZN7hipsycl2rt7ocl_usm20from_intel_extensionEPNS0_20ocl_hardware_managerEi(ptr dead_on_unwind nonnull writable sret(%"class.std::unique_ptr") align 8 %3, ptr noundef %1, i32 noundef %11)
  %12 = getelementptr inbounds i8, ptr %0, i64 40
  %13 = invoke noundef nonnull align 8 dereferenceable(16) ptr @_ZNSt10shared_ptrIN7hipsycl2rt7ocl_usmEEaSIS2_St14default_deleteIS2_EEENSt9enable_ifIXsr13is_assignableIRSt12__shared_ptrIS2_LN9__gnu_cxx12_Lock_policyE2EESt10unique_ptrIT_T0_EEE5valueERS3_E4typeEOSG_(ptr noundef nonnull align 8 dereferenceable(16) %12, ptr noundef nonnull align 8 dereferenceable(8) %3)
          to label %14 unwind label %33

14:                                               ; preds = %2
  call void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8) %3) #6
  call void @llvm.lifetime.end.p0(i64 8, ptr nonnull %3) #6
  %15 = call noundef ptr @_ZNKSt19__shared_ptr_accessIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2ELb0ELb0EEptEv(ptr noundef nonnull align 1 dereferenceable(1) %12) #6
  %16 = load ptr, ptr %15, align 8, !tbaa !27
  %17 = getelementptr inbounds i8, ptr %16, i64 16
  %18 = load ptr, ptr %17, align 8
  %19 = call noundef zeroext i1 %18(ptr noundef nonnull align 8 dereferenceable(8) %15)
  br i1 %19, label %53, label %20

20:                                               ; preds = %14
  call void @llvm.lifetime.start.p0(i64 8, ptr nonnull %4) #6
  %21 = load i32, ptr %10, align 8, !tbaa !5
  call void @_ZN7hipsycl2rt7ocl_usm28from_fine_grained_system_svmEPNS0_20ocl_hardware_managerEi(ptr dead_on_unwind nonnull writable sret(%"class.std::unique_ptr") align 8 %4, ptr noundef %1, i32 noundef %21)
  %22 = invoke noundef nonnull align 8 dereferenceable(16) ptr @_ZNSt10shared_ptrIN7hipsycl2rt7ocl_usmEEaSIS2_St14default_deleteIS2_EEENSt9enable_ifIXsr13is_assignableIRSt12__shared_ptrIS2_LN9__gnu_cxx12_Lock_policyE2EESt10unique_ptrIT_T0_EEE5valueERS3_E4typeEOSG_(ptr noundef nonnull align 8 dereferenceable(16) %12, ptr noundef nonnull align 8 dereferenceable(8) %4)
          to label %23 unwind label %35

23:                                               ; preds = %20
  call void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8) %4) #6
  call void @llvm.lifetime.end.p0(i64 8, ptr nonnull %4) #6
  %24 = call noundef ptr @_ZNKSt19__shared_ptr_accessIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2ELb0ELb0EEptEv(ptr noundef nonnull align 1 dereferenceable(1) %12) #6
  %25 = load ptr, ptr %24, align 8, !tbaa !27
  %26 = getelementptr inbounds i8, ptr %25, i64 16
  %27 = load ptr, ptr %26, align 8
  %28 = call noundef zeroext i1 %27(ptr noundef nonnull align 8 dereferenceable(8) %24)
  br i1 %28, label %29, label %53

29:                                               ; preds = %23
  %30 = call noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv()
  %31 = call noundef i32 @_ZNK7hipsycl6common13output_stream15get_debug_levelEv(ptr noundef nonnull align 8 dereferenceable(16) %30)
  %32 = icmp slt i32 %31, 2
  br i1 %32, label %53, label %37

33:                                               ; preds = %2
  %34 = landingpad { ptr, i32 }
          cleanup
  call void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8) %3) #6
  call void @llvm.lifetime.end.p0(i64 8, ptr nonnull %3) #6
  br label %84

35:                                               ; preds = %20
  %36 = landingpad { ptr, i32 }
          cleanup
  call void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8) %4) #6
  call void @llvm.lifetime.end.p0(i64 8, ptr nonnull %4) #6
  br label %84

37:                                               ; preds = %29
  %38 = call noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv()
  %39 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZNK7hipsycl6common13output_stream10get_streamEv(ptr noundef nonnull align 8 dereferenceable(16) %38)
  %40 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %39, ptr noundef nonnull @.str.7)
  %41 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %40, ptr noundef nonnull @.str.8)
  call void @llvm.lifetime.start.p0(i64 32, ptr nonnull %5) #6
  %42 = load ptr, ptr %0, align 8, !tbaa !27
  %43 = getelementptr inbounds i8, ptr %42, i64 32
  %44 = load ptr, ptr %43, align 8
  call void %44(ptr dead_on_unwind nonnull writable sret(%"class.std::__cxx11::basic_string") align 8 %5, ptr noundef nonnull align 8 dereferenceable(89) %0)
  %45 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsIcSt11char_traitsIcESaIcEERSt13basic_ostreamIT_T0_ES7_RKNSt7__cxx1112basic_stringIS4_S5_T1_EE(ptr noundef nonnull align 8 dereferenceable(8) %41, ptr noundef nonnull align 8 dereferenceable(32) %5)
          to label %46 unwind label %51

46:                                               ; preds = %37
  %47 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %45, ptr noundef nonnull @.str.9)
          to label %48 unwind label %51

48:                                               ; preds = %46
  %49 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZNSolsEPFRSoS_E(ptr noundef nonnull align 8 dereferenceable(8) %47, ptr noundef nonnull @_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_)
          to label %50 unwind label %51

50:                                               ; preds = %48
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %5) #6
  call void @llvm.lifetime.end.p0(i64 32, ptr nonnull %5) #6
  br label %53

51:                                               ; preds = %48, %46, %37
  %52 = landingpad { ptr, i32 }
          cleanup
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %5) #6
  call void @llvm.lifetime.end.p0(i64 32, ptr nonnull %5) #6
  br label %84

53:                                               ; preds = %23, %29, %50, %14
  %54 = call noundef ptr @_ZNKSt19__shared_ptr_accessIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2ELb0ELb0EEptEv(ptr noundef nonnull align 1 dereferenceable(1) %12) #6
  %55 = load ptr, ptr %54, align 8, !tbaa !27
  %56 = getelementptr inbounds i8, ptr %55, i64 16
  %57 = load ptr, ptr %56, align 8
  %58 = call noundef zeroext i1 %57(ptr noundef nonnull align 8 dereferenceable(8) %54)
  br i1 %58, label %79, label %59

59:                                               ; preds = %53
  %60 = call noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv()
  %61 = call noundef i32 @_ZNK7hipsycl6common13output_stream15get_debug_levelEv(ptr noundef nonnull align 8 dereferenceable(16) %60)
  %62 = icmp slt i32 %61, 2
  br i1 %62, label %79, label %63

63:                                               ; preds = %59
  %64 = call noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv()
  %65 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZNK7hipsycl6common13output_stream10get_streamEv(ptr noundef nonnull align 8 dereferenceable(16) %64)
  %66 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %65, ptr noundef nonnull @.str.7)
  %67 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %66, ptr noundef nonnull @.str.8)
  call void @llvm.lifetime.start.p0(i64 32, ptr nonnull %6) #6
  %68 = load ptr, ptr %0, align 8, !tbaa !27
  %69 = getelementptr inbounds i8, ptr %68, i64 32
  %70 = load ptr, ptr %69, align 8
  call void %70(ptr dead_on_unwind nonnull writable sret(%"class.std::__cxx11::basic_string") align 8 %6, ptr noundef nonnull align 8 dereferenceable(89) %0)
  %71 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsIcSt11char_traitsIcESaIcEERSt13basic_ostreamIT_T0_ES7_RKNSt7__cxx1112basic_stringIS4_S5_T1_EE(ptr noundef nonnull align 8 dereferenceable(8) %67, ptr noundef nonnull align 8 dereferenceable(32) %6)
          to label %72 unwind label %77

72:                                               ; preds = %63
  %73 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8) %71, ptr noundef nonnull @.str.10)
          to label %74 unwind label %77

74:                                               ; preds = %72
  %75 = invoke noundef nonnull align 8 dereferenceable(8) ptr @_ZNSolsEPFRSoS_E(ptr noundef nonnull align 8 dereferenceable(8) %73, ptr noundef nonnull @_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_)
          to label %76 unwind label %77

76:                                               ; preds = %74
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %6) #6
  call void @llvm.lifetime.end.p0(i64 32, ptr nonnull %6) #6
  br label %79

77:                                               ; preds = %74, %72, %63
  %78 = landingpad { ptr, i32 }
          cleanup
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %6) #6
  call void @llvm.lifetime.end.p0(i64 32, ptr nonnull %6) #6
  br label %84

79:                                               ; preds = %76, %59, %53
  call void @llvm.lifetime.start.p0(i64 16, ptr nonnull %7) #6
  call void @_ZN7hipsycl2rt18backend_descriptorC2ENS0_17hardware_platformENS0_12api_platformE(ptr noundef nonnull align 4 dereferenceable(12) %8, i32 noundef 3, i32 noundef 3)
  %80 = load i32, ptr %10, align 8, !tbaa !5
  %.sroa.01.0.copyload = load i64, ptr %8, align 8
  %.sroa.22.0..sroa_idx = getelementptr inbounds i8, ptr %8, i64 8
  %.sroa.22.0.copyload = load i32, ptr %.sroa.22.0..sroa_idx, align 8
  call void @_ZN7hipsycl2rt9device_idC1ENS0_18backend_descriptorEi(ptr noundef nonnull align 4 dereferenceable(16) %7, i64 %.sroa.01.0.copyload, i32 %.sroa.22.0.copyload, i32 noundef %80)
  call void @llvm.lifetime.start.p0(i64 32, ptr nonnull %9) #6
  %.sroa.0.0.copyload = load i64, ptr %7, align 8
  %.sroa.2.0..sroa_idx = getelementptr inbounds i8, ptr %7, i64 8
  %.sroa.2.0.copyload = load i64, ptr %.sroa.2.0..sroa_idx, align 8
  %81 = call noundef ptr @_ZNKSt12__shared_ptrIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2EE3getEv(ptr noundef nonnull align 8 dereferenceable(16) %12) #6
  call void @_ZN7hipsycl2rt13ocl_allocatorC1ENS0_9device_idEPNS0_7ocl_usmE(ptr noundef nonnull align 8 dereferenceable(32) %9, i64 %.sroa.0.0.copyload, i64 %.sroa.2.0.copyload, ptr noundef %81)
  %82 = getelementptr inbounds i8, ptr %0, i64 56
  %83 = call noundef nonnull align 8 dereferenceable(32) ptr @_ZN7hipsycl2rt13ocl_allocatoraSEOS1_(ptr noundef nonnull align 8 dereferenceable(32) %82, ptr noundef nonnull align 8 dereferenceable(32) %9) #6
  call void @_ZN7hipsycl2rt17backend_allocatorD2Ev(ptr noundef nonnull align 8 dereferenceable(32) %9) #6
  call void @llvm.lifetime.end.p0(i64 32, ptr nonnull %9) #6
  call void @llvm.lifetime.end.p0(i64 16, ptr nonnull %7) #6
  ret void

84:                                               ; preds = %77, %51, %35, %33
  %.pn = phi { ptr, i32 } [ %78, %77 ], [ %52, %51 ], [ %36, %35 ], [ %34, %33 ]
  resume { ptr, i32 } %.pn
}

declare void @_ZN7hipsycl2rt7ocl_usm20from_intel_extensionEPNS0_20ocl_hardware_managerEi(ptr dead_on_unwind writable sret(%"class.std::unique_ptr") align 8, ptr noundef, i32 noundef) local_unnamed_addr #3

; Function Attrs: mustprogress uwtable
declare dso_local noundef nonnull align 8 dereferenceable(16) ptr @_ZNSt10shared_ptrIN7hipsycl2rt7ocl_usmEEaSIS2_St14default_deleteIS2_EEENSt9enable_ifIXsr13is_assignableIRSt12__shared_ptrIS2_LN9__gnu_cxx12_Lock_policyE2EESt10unique_ptrIT_T0_EEE5valueERS3_E4typeEOSG_(ptr noundef nonnull align 8 dereferenceable(16), ptr noundef nonnull align 8 dereferenceable(8)) local_unnamed_addr #2 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local void @_ZNSt10unique_ptrIN7hipsycl2rt7ocl_usmESt14default_deleteIS2_EED2Ev(ptr noundef nonnull align 8 dereferenceable(8)) unnamed_addr #1 align 2

declare void @_ZN7hipsycl2rt7ocl_usm28from_fine_grained_system_svmEPNS0_20ocl_hardware_managerEi(ptr dead_on_unwind writable sret(%"class.std::unique_ptr") align 8, ptr noundef, i32 noundef) local_unnamed_addr #3

; Function Attrs: mustprogress uwtable
declare dso_local noundef nonnull align 8 dereferenceable(16) ptr @_ZN7hipsycl6common13output_stream3getEv() local_unnamed_addr #2 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef i32 @_ZNK7hipsycl6common13output_stream15get_debug_levelEv(ptr noundef nonnull align 8 dereferenceable(16)) local_unnamed_addr #1 align 2

; Function Attrs: inlinehint mustprogress uwtable
declare noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsISt11char_traitsIcEERSt13basic_ostreamIcT_ES5_PKc(ptr noundef nonnull align 8 dereferenceable(8), ptr noundef) local_unnamed_addr #4

; Function Attrs: inlinehint mustprogress uwtable
declare noundef nonnull align 8 dereferenceable(8) ptr @_ZStlsIcSt11char_traitsIcESaIcEERSt13basic_ostreamIT_T0_ES7_RKNSt7__cxx1112basic_stringIS4_S5_T1_EE(ptr noundef nonnull align 8 dereferenceable(8), ptr noundef nonnull align 8 dereferenceable(32)) local_unnamed_addr #4

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef nonnull align 8 dereferenceable(8) ptr @_ZNK7hipsycl6common13output_stream10get_streamEv(ptr noundef nonnull align 8 dereferenceable(16)) local_unnamed_addr #1 align 2

; Function Attrs: mustprogress uwtable
declare noundef nonnull align 8 dereferenceable(8) ptr @_ZNSolsEPFRSoS_E(ptr noundef nonnull align 8 dereferenceable(8), ptr noundef) local_unnamed_addr #2 align 2

; Function Attrs: inlinehint mustprogress uwtable
declare noundef nonnull align 8 dereferenceable(8) ptr @_ZSt4endlIcSt11char_traitsIcEERSt13basic_ostreamIT_T0_ES6_(ptr noundef nonnull align 8 dereferenceable(8)) #4

; Function Attrs: mustprogress nounwind uwtable
declare dso_local void @_ZN7hipsycl2rt18backend_descriptorC2ENS0_17hardware_platformENS0_12api_platformE(ptr noundef nonnull align 4 dereferenceable(12), i32 noundef, i32 noundef) unnamed_addr #1 align 2

declare void @_ZN7hipsycl2rt9device_idC1ENS0_18backend_descriptorEi(ptr noundef nonnull align 4 dereferenceable(16), i64, i32, i32 noundef) unnamed_addr #3

declare void @_ZN7hipsycl2rt13ocl_allocatorC1ENS0_9device_idEPNS0_7ocl_usmE(ptr noundef nonnull align 8 dereferenceable(32), i64, i64, ptr noundef) unnamed_addr #3

; Function Attrs: inlinehint mustprogress nounwind uwtable
declare dso_local noundef nonnull align 8 dereferenceable(32) ptr @_ZN7hipsycl2rt13ocl_allocatoraSEOS1_(ptr noundef nonnull align 8 dereferenceable(32), ptr noundef nonnull align 8 dereferenceable(32)) local_unnamed_addr #5 align 2

attributes #0 = { mustprogress nocallback nofree nosync nounwind willreturn memory(argmem: readwrite) }
attributes #1 = { mustprogress nounwind uwtable "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #2 = { mustprogress uwtable "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #3 = { "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #4 = { inlinehint mustprogress uwtable "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #5 = { inlinehint mustprogress nounwind uwtable "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "target-cpu"="x86-64" "target-features"="+cmov,+cx8,+fxsr,+mmx,+sse,+sse2,+x87" "tune-cpu"="generic" }
attributes #6 = { nounwind }

!llvm.module.flags = !{!0, !1, !2, !3}
!llvm.ident = !{!4}

!0 = !{i32 1, !"wchar_size", i32 4}
!1 = !{i32 8, !"PIC Level", i32 2}
!2 = !{i32 7, !"PIE Level", i32 2}
!3 = !{i32 7, !"uwtable", i32 2}
!4 = !{!"clang version 19.1.0 (git@github.com:llvm/llvm-project.git a4bf6cd7cfb1a1421ba92bca9d017b49936c55e4)"}
!5 = !{!6, !8, i64 8}
!6 = !{!"_ZTSN7hipsycl2rt20ocl_hardware_contextE", !7, i64 0, !8, i64 8, !8, i64 12, !11, i64 16, !14, i64 24, !17, i64 40, !20, i64 56, !16, i64 88}
!7 = !{!"_ZTSN7hipsycl2rt16hardware_contextE"}
!8 = !{!"int", !9, i64 0}
!9 = !{!"omnipotent char", !10, i64 0}
!10 = !{!"Simple C++ TBAA"}
!11 = !{!"_ZTSN2cl7ContextE", !12, i64 0}
!12 = !{!"_ZTSN2cl6detail7WrapperIP11_cl_contextEE", !13, i64 0}
!13 = !{!"any pointer", !9, i64 0}
!14 = !{!"_ZTSN2cl6DeviceE", !15, i64 0}
!15 = !{!"_ZTSN2cl6detail7WrapperIP13_cl_device_idEE", !13, i64 0, !16, i64 8}
!16 = !{!"bool", !9, i64 0}
!17 = !{!"_ZTSSt10shared_ptrIN7hipsycl2rt7ocl_usmEE", !18, i64 0}
!18 = !{!"_ZTSSt12__shared_ptrIN7hipsycl2rt7ocl_usmELN9__gnu_cxx12_Lock_policyE2EE", !13, i64 0, !19, i64 8}
!19 = !{!"_ZTSSt14__shared_countILN9__gnu_cxx12_Lock_policyE2EE", !13, i64 0}
!20 = !{!"_ZTSN7hipsycl2rt13ocl_allocatorE", !21, i64 0, !13, i64 8, !22, i64 16}
!21 = !{!"_ZTSN7hipsycl2rt17backend_allocatorE"}
!22 = !{!"_ZTSN7hipsycl2rt9device_idE", !23, i64 0, !8, i64 12}
!23 = !{!"_ZTSN7hipsycl2rt18backend_descriptorE", !24, i64 0, !25, i64 4, !26, i64 8}
!24 = !{!"_ZTSN7hipsycl2rt10backend_idE", !9, i64 0}
!25 = !{!"_ZTSN7hipsycl2rt17hardware_platformE", !9, i64 0}
!26 = !{!"_ZTSN7hipsycl2rt12api_platformE", !9, i64 0}
!27 = !{!28, !28, i64 0}
!28 = !{!"vtable pointer", !10, i64 0}


; Effective O3 passes: EarlyCSEPass,GlobalOptPass,InferAlignmentPass,InferFunctionAttrsPass,InstCombinePass,SROAPass,SimplifyCFGPass