; LLVM IR from AdaptiveCpp/AdaptiveCpp :: src/runtime/pcuda (pcuda_runtime_api)

; 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.anon = type { i8 }
%"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.std::vector.11" = type { %"struct.std::_Vector_base.12" }
%"struct.std::_Vector_base.12" = type { %"struct.std::_Vector_base<unsigned long, std::allocator<unsigned long>>::_Vector_impl" }
%"struct.std::_Vector_base<unsigned long, std::allocator<unsigned long>>::_Vector_impl" = type { %"struct.std::_Vector_base<unsigned long, std::allocator<unsigned long>>::_Vector_impl_data" }
%"struct.std::_Vector_base<unsigned long, std::allocator<unsigned long>>::_Vector_impl_data" = type { ptr, ptr, ptr }
%"struct.hipsycl::rt::pcuda::device_topology::device" = type { i32, %"class.hipsycl::rt::device_id", ptr }
%"class.hipsycl::rt::device_id" = type { %"struct.hipsycl::rt::backend_descriptor", i32 }
%"struct.hipsycl::rt::backend_descriptor" = type { i32, i32, i32 }
%struct.pcudaDeviceProp = type { [256 x i8], i64, i64, i32, i32, i64, i32, [3 x i32], [3 x i32], i32, i64, i32, i32, i64, i32, i32, i32, i32, i32, i32, i32, [2 x i32], [3 x i32], [2 x i32], [3 x i32], i64, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32, i32 }

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

declare noundef nonnull align 8 dereferenceable(104) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application3getEv() #1

declare noundef nonnull align 8 dereferenceable(96) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application9tls_stateEv(ptr noundef nonnull align 8 dereferenceable(104)) #1

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

declare noundef i32 @_ZN7hipsycl2rt5pcuda27get_most_recent_pcuda_errorEv() #1

declare i32 @__gxx_personality_v0(...)

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

declare noundef i32 @_ZNK7hipsycl2rt5pcuda18thread_local_state11get_backendEv(ptr noundef nonnull align 8 dereferenceable(96)) #1

declare noundef i32 @_ZNK7hipsycl2rt5pcuda18thread_local_state12get_platformEv(ptr noundef nonnull align 8 dereferenceable(96)) #1

declare noundef nonnull align 8 dereferenceable(40) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application8pcuda_rtEv(ptr noundef nonnull align 8 dereferenceable(104)) #1

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef nonnull align 8 dereferenceable(24) ptr @_ZNK7hipsycl2rt5pcuda13pcuda_runtime12get_topologyEv(ptr noundef nonnull align 8 dereferenceable(40)) #2 align 2

; Function Attrs: mustprogress uwtable
define dso_local i32 @pcudaGetDeviceProperties(ptr noundef %0, i32 noundef %1) #3 personality ptr @__gxx_personality_v0 {
  %3 = alloca i32, align 4
  %4 = alloca ptr, align 8
  %5 = alloca i32, align 4
  %6 = alloca i32, align 4
  %7 = alloca i32, align 4
  %8 = alloca i32, align 4
  %9 = alloca i32, align 4
  %10 = alloca ptr, align 8
  %11 = alloca %class.anon, align 1
  %12 = alloca ptr, align 8
  %13 = alloca %"class.std::__cxx11::basic_string", align 8
  %14 = alloca ptr, align 8
  %15 = alloca i32, align 4
  %16 = alloca %"class.std::vector.11", align 8
  store ptr %0, ptr %4, align 8, !tbaa !5
  store i32 %1, ptr %5, align 4, !tbaa !9
  call void @llvm.lifetime.start.p0(i64 4, ptr %6) #5
  %17 = call noundef i32 @_ZN7hipsycl2rt5pcuda27get_most_recent_pcuda_errorEv()
  store i32 %17, ptr %6, align 4, !tbaa !11
  %18 = load i32, ptr %6, align 4, !tbaa !11
  %19 = icmp ne i32 %18, 0
  br i1 %19, label %20, label %22

20:                                               ; preds = %2
  %21 = load i32, ptr %6, align 4, !tbaa !11
  store i32 %21, ptr %3, align 4
  store i32 1, ptr %7, align 4
  br label %247

22:                                               ; preds = %2
  call void @llvm.lifetime.start.p0(i64 4, ptr %8) #5
  %23 = call noundef nonnull align 8 dereferenceable(104) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application3getEv()
  %24 = call noundef nonnull align 8 dereferenceable(96) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application9tls_stateEv(ptr noundef nonnull align 8 dereferenceable(104) %23)
  %25 = call noundef i32 @_ZNK7hipsycl2rt5pcuda18thread_local_state11get_backendEv(ptr noundef nonnull align 8 dereferenceable(96) %24)
  store i32 %25, ptr %8, align 4, !tbaa !9
  call void @llvm.lifetime.start.p0(i64 4, ptr %9) #5
  %26 = call noundef nonnull align 8 dereferenceable(104) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application3getEv()
  %27 = call noundef nonnull align 8 dereferenceable(96) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application9tls_stateEv(ptr noundef nonnull align 8 dereferenceable(104) %26)
  %28 = call noundef i32 @_ZNK7hipsycl2rt5pcuda18thread_local_state12get_platformEv(ptr noundef nonnull align 8 dereferenceable(96) %27)
  store i32 %28, ptr %9, align 4, !tbaa !9
  call void @llvm.lifetime.start.p0(i64 8, ptr %10) #5
  %29 = call noundef nonnull align 8 dereferenceable(104) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application3getEv()
  %30 = call noundef nonnull align 8 dereferenceable(40) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application8pcuda_rtEv(ptr noundef nonnull align 8 dereferenceable(104) %29)
  %31 = call noundef nonnull align 8 dereferenceable(24) ptr @_ZNK7hipsycl2rt5pcuda13pcuda_runtime12get_topologyEv(ptr noundef nonnull align 8 dereferenceable(40) %30)
  %32 = load i32, ptr %8, align 4, !tbaa !9
  %33 = load i32, ptr %9, align 4, !tbaa !9
  %34 = load i32, ptr %5, align 4, !tbaa !9
  %35 = call noundef ptr @_ZNK7hipsycl2rt5pcuda15device_topology10get_deviceEiii(ptr noundef nonnull align 8 dereferenceable(24) %31, i32 noundef %32, i32 noundef %33, i32 noundef %34)
  store ptr %35, ptr %10, align 8, !tbaa !5
  %36 = load ptr, ptr %10, align 8, !tbaa !5
  %37 = icmp ne ptr %36, null
  br i1 %37, label %39, label %38

38:                                               ; preds = %22
  store i32 10, ptr %3, align 4
  store i32 1, ptr %7, align 4
  br label %246

39:                                               ; preds = %22
  call void @llvm.lifetime.start.p0(i64 1, ptr %11) #5
  call void @llvm.lifetime.start.p0(i64 8, ptr %12) #5
  %40 = load ptr, ptr %10, align 8, !tbaa !5
  %41 = getelementptr inbounds %"struct.hipsycl::rt::pcuda::device_topology::device", ptr %40, i32 0, i32 2
  %42 = load ptr, ptr %41, align 8, !tbaa !13
  store ptr %42, ptr %12, align 8, !tbaa !5
  call void @llvm.lifetime.start.p0(i64 32, ptr %13) #5
  %43 = load ptr, ptr %12, align 8, !tbaa !5
  %44 = load ptr, ptr %43, align 8, !tbaa !20
  %45 = getelementptr inbounds ptr, ptr %44, i64 4
  %46 = load ptr, ptr %45, align 8
  call void %46(ptr dead_on_unwind writable sret(%"class.std::__cxx11::basic_string") align 8 %13, ptr noundef nonnull align 8 dereferenceable(8) %43)
  %47 = load ptr, ptr %4, align 8, !tbaa !5
  %48 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %47, i32 0, i32 0
  %49 = getelementptr inbounds [256 x i8], ptr %48, i64 0, i64 0
  invoke void @"_ZZ24pcudaGetDevicePropertiesENK3$_0clERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEEmPc"(ptr noundef nonnull align 1 dereferenceable(1) %11, ptr noundef nonnull align 8 dereferenceable(32) %13, i64 noundef 256, ptr noundef %49)
          to label %50 unwind label %242

50:                                               ; preds = %39
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %13) #5
  call void @llvm.lifetime.end.p0(i64 32, ptr %13) #5
  %51 = load ptr, ptr %12, align 8, !tbaa !5
  %52 = load ptr, ptr %51, align 8, !tbaa !20
  %53 = getelementptr inbounds ptr, ptr %52, i64 8
  %54 = load ptr, ptr %53, align 8
  %55 = call noundef i64 %54(ptr noundef nonnull align 8 dereferenceable(8) %51, i32 noundef 41)
  %56 = load ptr, ptr %4, align 8, !tbaa !5
  %57 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %56, i32 0, i32 1
  store i64 %55, ptr %57, align 8, !tbaa !22
  %58 = load ptr, ptr %12, align 8, !tbaa !5
  %59 = load ptr, ptr %58, align 8, !tbaa !20
  %60 = getelementptr inbounds ptr, ptr %59, i64 8
  %61 = load ptr, ptr %60, align 8
  %62 = call noundef i64 %61(ptr noundef nonnull align 8 dereferenceable(8) %58, i32 noundef 44)
  %63 = load ptr, ptr %4, align 8, !tbaa !5
  %64 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %63, i32 0, i32 2
  store i64 %62, ptr %64, align 8, !tbaa !25
  %65 = load ptr, ptr %4, align 8, !tbaa !5
  %66 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %65, i32 0, i32 3
  store i32 0, ptr %66, align 8, !tbaa !26
  call void @llvm.lifetime.start.p0(i64 24, ptr %16) #5
  %67 = load ptr, ptr %12, align 8, !tbaa !5
  %68 = load ptr, ptr %67, align 8, !tbaa !20
  %69 = getelementptr inbounds ptr, ptr %68, i64 9
  %70 = load ptr, ptr %69, align 8
  call void %70(ptr dead_on_unwind writable sret(%"class.std::vector.11") align 8 %16, ptr noundef nonnull align 8 dereferenceable(8) %67, i32 noundef 0)
  %71 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZNSt6vectorImSaImEEixEm(ptr noundef nonnull align 8 dereferenceable(24) %16, i64 noundef 0) #5
  %72 = load i64, ptr %71, align 8, !tbaa !27
  %73 = trunc i64 %72 to i32
  %74 = load ptr, ptr %4, align 8, !tbaa !5
  %75 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %74, i32 0, i32 4
  store i32 %73, ptr %75, align 4, !tbaa !28
  call void @_ZNSt6vectorImSaImEED2Ev(ptr noundef nonnull align 8 dereferenceable(24) %16) #5
  call void @llvm.lifetime.end.p0(i64 24, ptr %16) #5
  %76 = load ptr, ptr %4, align 8, !tbaa !5
  %77 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %76, i32 0, i32 5
  store i64 0, ptr %77, align 8, !tbaa !29
  %78 = load ptr, ptr %12, align 8, !tbaa !5
  %79 = load ptr, ptr %78, align 8, !tbaa !20
  %80 = getelementptr inbounds ptr, ptr %79, i64 8
  %81 = load ptr, ptr %80, align 8
  %82 = call noundef i64 %81(ptr noundef nonnull align 8 dereferenceable(8) %78, i32 noundef 8)
  %83 = trunc i64 %82 to i32
  %84 = load ptr, ptr %4, align 8, !tbaa !5
  %85 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %84, i32 0, i32 6
  store i32 %83, ptr %85, align 8, !tbaa !30
  %86 = load ptr, ptr %12, align 8, !tbaa !5
  %87 = load ptr, ptr %86, align 8, !tbaa !20
  %88 = getelementptr inbounds ptr, ptr %87, i64 8
  %89 = load ptr, ptr %88, align 8
  %90 = call noundef i64 %89(ptr noundef nonnull align 8 dereferenceable(8) %86, i32 noundef 5)
  %91 = trunc i64 %90 to i32
  %92 = load ptr, ptr %4, align 8, !tbaa !5
  %93 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %92, i32 0, i32 7
  %94 = getelementptr inbounds [3 x i32], ptr %93, i64 0, i64 0
  store i32 %91, ptr %94, align 4, !tbaa !9
  %95 = load ptr, ptr %12, align 8, !tbaa !5
  %96 = load ptr, ptr %95, align 8, !tbaa !20
  %97 = getelementptr inbounds ptr, ptr %96, i64 8
  %98 = load ptr, ptr %97, align 8
  %99 = call noundef i64 %98(ptr noundef nonnull align 8 dereferenceable(8) %95, i32 noundef 6)
  %100 = trunc i64 %99 to i32
  %101 = load ptr, ptr %4, align 8, !tbaa !5
  %102 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %101, i32 0, i32 7
  %103 = getelementptr inbounds [3 x i32], ptr %102, i64 0, i64 1
  store i32 %100, ptr %103, align 4, !tbaa !9
  %104 = load ptr, ptr %12, align 8, !tbaa !5
  %105 = load ptr, ptr %104, align 8, !tbaa !20
  %106 = getelementptr inbounds ptr, ptr %105, i64 8
  %107 = load ptr, ptr %106, align 8
  %108 = call noundef i64 %107(ptr noundef nonnull align 8 dereferenceable(8) %104, i32 noundef 7)
  %109 = trunc i64 %108 to i32
  %110 = load ptr, ptr %4, align 8, !tbaa !5
  %111 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %110, i32 0, i32 7
  %112 = getelementptr inbounds [3 x i32], ptr %111, i64 0, i64 2
  store i32 %109, ptr %112, align 4, !tbaa !9
  %113 = call noundef i32 @_ZNSt14numeric_limitsIiE3maxEv() #5
  %114 = load ptr, ptr %4, align 8, !tbaa !5
  %115 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %114, i32 0, i32 8
  %116 = getelementptr inbounds [3 x i32], ptr %115, i64 0, i64 0
  store i32 %113, ptr %116, align 8, !tbaa !9
  %117 = call noundef i32 @_ZNSt14numeric_limitsIiE3maxEv() #5
  %118 = load ptr, ptr %4, align 8, !tbaa !5
  %119 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %118, i32 0, i32 8
  %120 = getelementptr inbounds [3 x i32], ptr %119, i64 0, i64 1
  store i32 %117, ptr %120, align 4, !tbaa !9
  %121 = call noundef i32 @_ZNSt14numeric_limitsIiE3maxEv() #5
  %122 = load ptr, ptr %4, align 8, !tbaa !5
  %123 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %122, i32 0, i32 8
  %124 = getelementptr inbounds [3 x i32], ptr %123, i64 0, i64 2
  store i32 %121, ptr %124, align 8, !tbaa !9
  %125 = load ptr, ptr %12, align 8, !tbaa !5
  %126 = load ptr, ptr %125, align 8, !tbaa !20
  %127 = getelementptr inbounds ptr, ptr %126, i64 8
  %128 = load ptr, ptr %127, align 8
  %129 = call noundef i64 %128(ptr noundef nonnull align 8 dereferenceable(8) %125, i32 noundef 24)
  %130 = trunc i64 %129 to i32
  %131 = load ptr, ptr %4, align 8, !tbaa !5
  %132 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %131, i32 0, i32 9
  store i32 %130, ptr %132, align 4, !tbaa !31
  %133 = load ptr, ptr %4, align 8, !tbaa !5
  %134 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %133, i32 0, i32 10
  store i64 0, ptr %134, align 8, !tbaa !32
  %135 = load ptr, ptr %4, align 8, !tbaa !5
  %136 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %135, i32 0, i32 11
  store i32 0, ptr %136, align 8, !tbaa !33
  %137 = load ptr, ptr %4, align 8, !tbaa !5
  %138 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %137, i32 0, i32 12
  store i32 0, ptr %138, align 4, !tbaa !34
  %139 = load ptr, ptr %4, align 8, !tbaa !5
  %140 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %139, i32 0, i32 13
  store i64 0, ptr %140, align 8, !tbaa !35
  %141 = load ptr, ptr %4, align 8, !tbaa !5
  %142 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %141, i32 0, i32 14
  store i32 1, ptr %142, align 8, !tbaa !36
  %143 = load ptr, ptr %12, align 8, !tbaa !5
  %144 = load ptr, ptr %143, align 8, !tbaa !20
  %145 = getelementptr inbounds ptr, ptr %144, i64 8
  %146 = load ptr, ptr %145, align 8
  %147 = call noundef i64 %146(ptr noundef nonnull align 8 dereferenceable(8) %143, i32 noundef 0)
  %148 = trunc i64 %147 to i32
  %149 = load ptr, ptr %4, align 8, !tbaa !5
  %150 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %149, i32 0, i32 15
  store i32 %148, ptr %150, align 4, !tbaa !37
  %151 = load ptr, ptr %4, align 8, !tbaa !5
  %152 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %151, i32 0, i32 16
  store i32 0, ptr %152, align 8, !tbaa !38
  %153 = load ptr, ptr %4, align 8, !tbaa !5
  %154 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %153, i32 0, i32 17
  store i32 0, ptr %154, align 4, !tbaa !39
  %155 = load ptr, ptr %4, align 8, !tbaa !5
  %156 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %155, i32 0, i32 18
  store i32 1, ptr %156, align 8, !tbaa !40
  %157 = load ptr, ptr %4, align 8, !tbaa !5
  %158 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %157, i32 0, i32 19
  store i32 0, ptr %158, align 4, !tbaa !41
  %159 = load ptr, ptr %4, align 8, !tbaa !5
  %160 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %159, i32 0, i32 20
  store i32 0, ptr %160, align 8, !tbaa !42
  %161 = load ptr, ptr %4, align 8, !tbaa !5
  %162 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %161, i32 0, i32 21
  %163 = getelementptr inbounds [2 x i32], ptr %162, i64 0, i64 0
  store i32 0, ptr %163, align 4, !tbaa !9
  %164 = load ptr, ptr %4, align 8, !tbaa !5
  %165 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %164, i32 0, i32 21
  %166 = getelementptr inbounds [2 x i32], ptr %165, i64 0, i64 1
  store i32 0, ptr %166, align 4, !tbaa !9
  %167 = load ptr, ptr %4, align 8, !tbaa !5
  %168 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %167, i32 0, i32 22
  %169 = getelementptr inbounds [3 x i32], ptr %168, i64 0, i64 0
  store i32 0, ptr %169, align 4, !tbaa !9
  %170 = load ptr, ptr %4, align 8, !tbaa !5
  %171 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %170, i32 0, i32 22
  %172 = getelementptr inbounds [3 x i32], ptr %171, i64 0, i64 1
  store i32 0, ptr %172, align 4, !tbaa !9
  %173 = load ptr, ptr %4, align 8, !tbaa !5
  %174 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %173, i32 0, i32 22
  %175 = getelementptr inbounds [3 x i32], ptr %174, i64 0, i64 2
  store i32 0, ptr %175, align 4, !tbaa !9
  %176 = load ptr, ptr %4, align 8, !tbaa !5
  %177 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %176, i32 0, i32 23
  %178 = getelementptr inbounds [2 x i32], ptr %177, i64 0, i64 0
  store i32 0, ptr %178, align 8, !tbaa !9
  %179 = load ptr, ptr %4, align 8, !tbaa !5
  %180 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %179, i32 0, i32 23
  %181 = getelementptr inbounds [2 x i32], ptr %180, i64 0, i64 1
  store i32 0, ptr %181, align 4, !tbaa !9
  %182 = load ptr, ptr %4, align 8, !tbaa !5
  %183 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %182, i32 0, i32 24
  %184 = getelementptr inbounds [3 x i32], ptr %183, i64 0, i64 0
  store i32 0, ptr %184, align 8, !tbaa !9
  %185 = load ptr, ptr %4, align 8, !tbaa !5
  %186 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %185, i32 0, i32 24
  %187 = getelementptr inbounds [3 x i32], ptr %186, i64 0, i64 1
  store i32 0, ptr %187, align 4, !tbaa !9
  %188 = load ptr, ptr %4, align 8, !tbaa !5
  %189 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %188, i32 0, i32 24
  %190 = getelementptr inbounds [3 x i32], ptr %189, i64 0, i64 2
  store i32 0, ptr %190, align 8, !tbaa !9
  %191 = load ptr, ptr %4, align 8, !tbaa !5
  %192 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %191, i32 0, i32 25
  store i64 0, ptr %192, align 8, !tbaa !43
  %193 = load ptr, ptr %12, align 8, !tbaa !5
  %194 = load ptr, ptr %193, align 8, !tbaa !20
  %195 = getelementptr inbounds ptr, ptr %194, i64 2
  %196 = load ptr, ptr %195, align 8
  %197 = call noundef i64 %196(ptr noundef nonnull align 8 dereferenceable(8) %193)
  %198 = icmp ugt i64 %197, 1
  %199 = zext i1 %198 to i32
  %200 = load ptr, ptr %4, align 8, !tbaa !5
  %201 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %200, i32 0, i32 26
  store i32 %199, ptr %201, align 8, !tbaa !44
  %202 = load ptr, ptr %4, align 8, !tbaa !5
  %203 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %202, i32 0, i32 27
  store i32 0, ptr %203, align 4, !tbaa !45
  %204 = load ptr, ptr %4, align 8, !tbaa !5
  %205 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %204, i32 0, i32 28
  store i32 0, ptr %205, align 8, !tbaa !46
  %206 = load ptr, ptr %4, align 8, !tbaa !5
  %207 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %206, i32 0, i32 29
  store i32 0, ptr %207, align 4, !tbaa !47
  %208 = load ptr, ptr %4, align 8, !tbaa !5
  %209 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %208, i32 0, i32 30
  store i32 0, ptr %209, align 8, !tbaa !48
  %210 = load ptr, ptr %4, align 8, !tbaa !5
  %211 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %210, i32 0, i32 31
  store i32 0, ptr %211, align 4, !tbaa !49
  %212 = load ptr, ptr %12, align 8, !tbaa !5
  %213 = load ptr, ptr %212, align 8, !tbaa !20
  %214 = getelementptr inbounds ptr, ptr %213, i64 3
  %215 = load ptr, ptr %214, align 8
  %216 = call noundef i64 %215(ptr noundef nonnull align 8 dereferenceable(8) %212)
  %217 = trunc i64 %216 to i32
  %218 = load ptr, ptr %4, align 8, !tbaa !5
  %219 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %218, i32 0, i32 32
  store i32 %217, ptr %219, align 8, !tbaa !50
  %220 = load ptr, ptr %4, align 8, !tbaa !5
  %221 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %220, i32 0, i32 33
  store i32 1, ptr %221, align 4, !tbaa !51
  %222 = load ptr, ptr %4, align 8, !tbaa !5
  %223 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %222, i32 0, i32 34
  store i32 0, ptr %223, align 8, !tbaa !52
  %224 = load ptr, ptr %4, align 8, !tbaa !5
  %225 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %224, i32 0, i32 35
  store i32 0, ptr %225, align 4, !tbaa !53
  %226 = load ptr, ptr %12, align 8, !tbaa !5
  %227 = load ptr, ptr %226, align 8, !tbaa !20
  %228 = getelementptr inbounds ptr, ptr %227, i64 8
  %229 = load ptr, ptr %228, align 8
  %230 = call noundef i64 %229(ptr noundef nonnull align 8 dereferenceable(8) %226, i32 noundef 40)
  %231 = trunc i64 %230 to i32
  %232 = load ptr, ptr %4, align 8, !tbaa !5
  %233 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %232, i32 0, i32 36
  store i32 %231, ptr %233, align 8, !tbaa !54
  %234 = load ptr, ptr %12, align 8, !tbaa !5
  %235 = load ptr, ptr %234, align 8, !tbaa !20
  %236 = getelementptr inbounds ptr, ptr %235, i64 8
  %237 = load ptr, ptr %236, align 8
  %238 = call noundef i64 %237(ptr noundef nonnull align 8 dereferenceable(8) %234, i32 noundef 8)
  %239 = trunc i64 %238 to i32
  %240 = load ptr, ptr %4, align 8, !tbaa !5
  %241 = getelementptr inbounds %struct.pcudaDeviceProp, ptr %240, i32 0, i32 37
  store i32 %239, ptr %241, align 4, !tbaa !55
  store i32 0, ptr %3, align 4
  store i32 1, ptr %7, align 4
  call void @llvm.lifetime.end.p0(i64 8, ptr %12) #5
  call void @llvm.lifetime.end.p0(i64 1, ptr %11) #5
  br label %246

242:                                              ; preds = %39
  %243 = landingpad { ptr, i32 }
          cleanup
  %244 = extractvalue { ptr, i32 } %243, 0
  store ptr %244, ptr %14, align 8
  %245 = extractvalue { ptr, i32 } %243, 1
  store i32 %245, ptr %15, align 4
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %13) #5
  call void @llvm.lifetime.end.p0(i64 32, ptr %13) #5
  call void @llvm.lifetime.end.p0(i64 8, ptr %12) #5
  call void @llvm.lifetime.end.p0(i64 1, ptr %11) #5
  call void @llvm.lifetime.end.p0(i64 8, ptr %10) #5
  call void @llvm.lifetime.end.p0(i64 4, ptr %9) #5
  call void @llvm.lifetime.end.p0(i64 4, ptr %8) #5
  call void @llvm.lifetime.end.p0(i64 4, ptr %6) #5
  br label %249

246:                                              ; preds = %50, %38
  call void @llvm.lifetime.end.p0(i64 8, ptr %10) #5
  call void @llvm.lifetime.end.p0(i64 4, ptr %9) #5
  call void @llvm.lifetime.end.p0(i64 4, ptr %8) #5
  br label %247

247:                                              ; preds = %246, %20
  call void @llvm.lifetime.end.p0(i64 4, ptr %6) #5
  %248 = load i32, ptr %3, align 4
  ret i32 %248

249:                                              ; preds = %242
  %250 = load ptr, ptr %14, align 8
  %251 = load i32, ptr %15, align 4
  %252 = insertvalue { ptr, i32 } poison, ptr %250, 0
  %253 = insertvalue { ptr, i32 } %252, i32 %251, 1
  resume { ptr, i32 } %253
}

declare noundef ptr @_ZNK7hipsycl2rt5pcuda15device_topology10get_deviceEiii(ptr noundef nonnull align 8 dereferenceable(24), i32 noundef, i32 noundef, i32 noundef) #1

; Function Attrs: inlinehint mustprogress uwtable
declare hidden void @"_ZZ24pcudaGetDevicePropertiesENK3$_0clERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEEmPc"(ptr noundef nonnull align 1 dereferenceable(1), ptr noundef nonnull align 8 dereferenceable(32), i64 noundef, ptr noundef) #4 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef nonnull align 8 dereferenceable(8) ptr @_ZNSt6vectorImSaImEEixEm(ptr noundef nonnull align 8 dereferenceable(24), i64 noundef) #2 align 2

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

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef i32 @_ZNSt14numeric_limitsIiE3maxEv() #2 align 2

attributes #0 = { nocallback nofree nosync nounwind willreturn memory(argmem: readwrite) }
attributes #1 = { "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 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 #3 = { 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 #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 = { 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, !10, i64 0}
!10 = !{!"int", !7, i64 0}
!11 = !{!12, !12, i64 0}
!12 = !{!"_ZTS10pcudaError", !7, i64 0}
!13 = !{!14, !6, i64 24}
!14 = !{!"_ZTSN7hipsycl2rt5pcuda15device_topology6deviceE", !10, i64 0, !15, i64 4, !6, i64 24}
!15 = !{!"_ZTSN7hipsycl2rt9device_idE", !16, i64 0, !10, i64 12}
!16 = !{!"_ZTSN7hipsycl2rt18backend_descriptorE", !17, i64 0, !18, i64 4, !19, i64 8}
!17 = !{!"_ZTSN7hipsycl2rt10backend_idE", !7, i64 0}
!18 = !{!"_ZTSN7hipsycl2rt17hardware_platformE", !7, i64 0}
!19 = !{!"_ZTSN7hipsycl2rt12api_platformE", !7, i64 0}
!20 = !{!21, !21, i64 0}
!21 = !{!"vtable pointer", !8, i64 0}
!22 = !{!23, !24, i64 256}
!23 = !{!"_ZTS15pcudaDeviceProp", !7, i64 0, !24, i64 256, !24, i64 264, !10, i64 272, !10, i64 276, !24, i64 280, !10, i64 288, !7, i64 292, !7, i64 304, !10, i64 316, !24, i64 320, !10, i64 328, !10, i64 332, !24, i64 336, !10, i64 344, !10, i64 348, !10, i64 352, !10, i64 356, !10, i64 360, !10, i64 364, !10, i64 368, !7, i64 372, !7, i64 380, !7, i64 392, !7, i64 400, !24, i64 416, !10, i64 424, !10, i64 428, !10, i64 432, !10, i64 436, !10, i64 440, !10, i64 444, !10, i64 448, !10, i64 452, !10, i64 456, !10, i64 460, !10, i64 464, !10, i64 468}
!24 = !{!"long", !7, i64 0}
!25 = !{!23, !24, i64 264}
!26 = !{!23, !10, i64 272}
!27 = !{!24, !24, i64 0}
!28 = !{!23, !10, i64 276}
!29 = !{!23, !24, i64 280}
!30 = !{!23, !10, i64 288}
!31 = !{!23, !10, i64 316}
!32 = !{!23, !24, i64 320}
!33 = !{!23, !10, i64 328}
!34 = !{!23, !10, i64 332}
!35 = !{!23, !24, i64 336}
!36 = !{!23, !10, i64 344}
!37 = !{!23, !10, i64 348}
!38 = !{!23, !10, i64 352}
!39 = !{!23, !10, i64 356}
!40 = !{!23, !10, i64 360}
!41 = !{!23, !10, i64 364}
!42 = !{!23, !10, i64 368}
!43 = !{!23, !24, i64 416}
!44 = !{!23, !10, i64 424}
!45 = !{!23, !10, i64 428}
!46 = !{!23, !10, i64 432}
!47 = !{!23, !10, i64 436}
!48 = !{!23, !10, i64 440}
!49 = !{!23, !10, i64 444}
!50 = !{!23, !10, i64 448}
!51 = !{!23, !10, i64 452}
!52 = !{!23, !10, i64 456}
!53 = !{!23, !10, i64 460}
!54 = !{!23, !10, i64 464}
!55 = !{!23, !10, i64 468}


; 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.anon = type { i8 }
%"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.std::vector.11" = type { %"struct.std::_Vector_base.12" }
%"struct.std::_Vector_base.12" = type { %"struct.std::_Vector_base<unsigned long, std::allocator<unsigned long>>::_Vector_impl" }
%"struct.std::_Vector_base<unsigned long, std::allocator<unsigned long>>::_Vector_impl" = type { %"struct.std::_Vector_base<unsigned long, std::allocator<unsigned long>>::_Vector_impl_data" }
%"struct.std::_Vector_base<unsigned long, std::allocator<unsigned long>>::_Vector_impl_data" = type { ptr, ptr, ptr }

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

declare noundef nonnull align 8 dereferenceable(104) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application3getEv() local_unnamed_addr #1

declare noundef nonnull align 8 dereferenceable(96) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application9tls_stateEv(ptr noundef nonnull align 8 dereferenceable(104)) local_unnamed_addr #1

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

declare noundef i32 @_ZN7hipsycl2rt5pcuda27get_most_recent_pcuda_errorEv() local_unnamed_addr #1

declare i32 @__gxx_personality_v0(...)

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

declare noundef i32 @_ZNK7hipsycl2rt5pcuda18thread_local_state11get_backendEv(ptr noundef nonnull align 8 dereferenceable(96)) local_unnamed_addr #1

declare noundef i32 @_ZNK7hipsycl2rt5pcuda18thread_local_state12get_platformEv(ptr noundef nonnull align 8 dereferenceable(96)) local_unnamed_addr #1

declare noundef nonnull align 8 dereferenceable(40) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application8pcuda_rtEv(ptr noundef nonnull align 8 dereferenceable(104)) local_unnamed_addr #1

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef nonnull align 8 dereferenceable(24) ptr @_ZNK7hipsycl2rt5pcuda13pcuda_runtime12get_topologyEv(ptr noundef nonnull align 8 dereferenceable(40)) local_unnamed_addr #2 align 2

; Function Attrs: mustprogress uwtable
define dso_local noundef i32 @pcudaGetDeviceProperties(ptr noundef %0, i32 noundef %1) local_unnamed_addr #3 personality ptr @__gxx_personality_v0 {
  %3 = alloca %class.anon, align 1
  %4 = alloca %"class.std::__cxx11::basic_string", align 8
  %5 = alloca %"class.std::vector.11", align 8
  %6 = tail call noundef i32 @_ZN7hipsycl2rt5pcuda27get_most_recent_pcuda_errorEv()
  %.not = icmp eq i32 %6, 0
  br i1 %.not, label %7, label %124

7:                                                ; preds = %2
  %8 = tail call noundef nonnull align 8 dereferenceable(104) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application3getEv()
  %9 = tail call noundef nonnull align 8 dereferenceable(96) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application9tls_stateEv(ptr noundef nonnull align 8 dereferenceable(104) %8)
  %10 = tail call noundef i32 @_ZNK7hipsycl2rt5pcuda18thread_local_state11get_backendEv(ptr noundef nonnull align 8 dereferenceable(96) %9)
  %11 = tail call noundef nonnull align 8 dereferenceable(104) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application3getEv()
  %12 = tail call noundef nonnull align 8 dereferenceable(96) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application9tls_stateEv(ptr noundef nonnull align 8 dereferenceable(104) %11)
  %13 = tail call noundef i32 @_ZNK7hipsycl2rt5pcuda18thread_local_state12get_platformEv(ptr noundef nonnull align 8 dereferenceable(96) %12)
  %14 = tail call noundef nonnull align 8 dereferenceable(104) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application3getEv()
  %15 = tail call noundef nonnull align 8 dereferenceable(40) ptr @_ZN7hipsycl2rt5pcuda17pcuda_application8pcuda_rtEv(ptr noundef nonnull align 8 dereferenceable(104) %14)
  %16 = tail call noundef nonnull align 8 dereferenceable(24) ptr @_ZNK7hipsycl2rt5pcuda13pcuda_runtime12get_topologyEv(ptr noundef nonnull align 8 dereferenceable(40) %15)
  %17 = tail call noundef ptr @_ZNK7hipsycl2rt5pcuda15device_topology10get_deviceEiii(ptr noundef nonnull align 8 dereferenceable(24) %16, i32 noundef %10, i32 noundef %13, i32 noundef %1)
  %.not73 = icmp eq ptr %17, null
  br i1 %.not73, label %124, label %18

18:                                               ; preds = %7
  call void @llvm.lifetime.start.p0(i64 1, ptr nonnull %3) #6
  %19 = getelementptr inbounds i8, ptr %17, i64 24
  %20 = load ptr, ptr %19, align 8, !tbaa !5
  call void @llvm.lifetime.start.p0(i64 32, ptr nonnull %4) #6
  %21 = load ptr, ptr %20, align 8, !tbaa !16
  %22 = getelementptr inbounds i8, ptr %21, i64 32
  %23 = load ptr, ptr %22, align 8
  call void %23(ptr dead_on_unwind nonnull writable sret(%"class.std::__cxx11::basic_string") align 8 %4, ptr noundef nonnull align 8 dereferenceable(8) %20)
  invoke void @"_ZZ24pcudaGetDevicePropertiesENK3$_0clERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEEmPc"(ptr noundef nonnull align 1 dereferenceable(1) %3, ptr noundef nonnull align 8 dereferenceable(32) %4, i64 noundef 256, ptr noundef %0)
          to label %24 unwind label %122

24:                                               ; preds = %18
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %4) #6
  call void @llvm.lifetime.end.p0(i64 32, ptr nonnull %4) #6
  %25 = load ptr, ptr %20, align 8, !tbaa !16
  %26 = getelementptr inbounds i8, ptr %25, i64 64
  %27 = load ptr, ptr %26, align 8
  %28 = call noundef i64 %27(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 41)
  %29 = getelementptr inbounds i8, ptr %0, i64 256
  store i64 %28, ptr %29, align 8, !tbaa !18
  %30 = load ptr, ptr %20, align 8, !tbaa !16
  %31 = getelementptr inbounds i8, ptr %30, i64 64
  %32 = load ptr, ptr %31, align 8
  %33 = call noundef i64 %32(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 44)
  %34 = getelementptr inbounds i8, ptr %0, i64 264
  store i64 %33, ptr %34, align 8, !tbaa !21
  %35 = getelementptr inbounds i8, ptr %0, i64 272
  store i32 0, ptr %35, align 8, !tbaa !22
  call void @llvm.lifetime.start.p0(i64 24, ptr nonnull %5) #6
  %36 = load ptr, ptr %20, align 8, !tbaa !16
  %37 = getelementptr inbounds i8, ptr %36, i64 72
  %38 = load ptr, ptr %37, align 8
  call void %38(ptr dead_on_unwind nonnull writable sret(%"class.std::vector.11") align 8 %5, ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 0)
  %39 = call noundef nonnull align 8 dereferenceable(8) ptr @_ZNSt6vectorImSaImEEixEm(ptr noundef nonnull align 8 dereferenceable(24) %5, i64 noundef 0) #6
  %40 = load i64, ptr %39, align 8, !tbaa !23
  %41 = trunc i64 %40 to i32
  %42 = getelementptr inbounds i8, ptr %0, i64 276
  store i32 %41, ptr %42, align 4, !tbaa !24
  call void @_ZNSt6vectorImSaImEED2Ev(ptr noundef nonnull align 8 dereferenceable(24) %5) #6
  call void @llvm.lifetime.end.p0(i64 24, ptr nonnull %5) #6
  %43 = getelementptr inbounds i8, ptr %0, i64 280
  store i64 0, ptr %43, align 8, !tbaa !25
  %44 = load ptr, ptr %20, align 8, !tbaa !16
  %45 = getelementptr inbounds i8, ptr %44, i64 64
  %46 = load ptr, ptr %45, align 8
  %47 = call noundef i64 %46(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 8)
  %48 = trunc i64 %47 to i32
  %49 = getelementptr inbounds i8, ptr %0, i64 288
  store i32 %48, ptr %49, align 8, !tbaa !26
  %50 = load ptr, ptr %20, align 8, !tbaa !16
  %51 = getelementptr inbounds i8, ptr %50, i64 64
  %52 = load ptr, ptr %51, align 8
  %53 = call noundef i64 %52(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 5)
  %54 = trunc i64 %53 to i32
  %55 = getelementptr inbounds i8, ptr %0, i64 292
  store i32 %54, ptr %55, align 4, !tbaa !27
  %56 = load ptr, ptr %20, align 8, !tbaa !16
  %57 = getelementptr inbounds i8, ptr %56, i64 64
  %58 = load ptr, ptr %57, align 8
  %59 = call noundef i64 %58(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 6)
  %60 = trunc i64 %59 to i32
  %61 = getelementptr inbounds i8, ptr %0, i64 296
  store i32 %60, ptr %61, align 4, !tbaa !27
  %62 = load ptr, ptr %20, align 8, !tbaa !16
  %63 = getelementptr inbounds i8, ptr %62, i64 64
  %64 = load ptr, ptr %63, align 8
  %65 = call noundef i64 %64(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 7)
  %66 = trunc i64 %65 to i32
  %67 = getelementptr inbounds i8, ptr %0, i64 300
  store i32 %66, ptr %67, align 4, !tbaa !27
  %68 = call noundef i32 @_ZNSt14numeric_limitsIiE3maxEv() #6
  %69 = getelementptr inbounds i8, ptr %0, i64 304
  store i32 %68, ptr %69, align 8, !tbaa !27
  %70 = call noundef i32 @_ZNSt14numeric_limitsIiE3maxEv() #6
  %71 = getelementptr inbounds i8, ptr %0, i64 308
  store i32 %70, ptr %71, align 4, !tbaa !27
  %72 = call noundef i32 @_ZNSt14numeric_limitsIiE3maxEv() #6
  %73 = getelementptr inbounds i8, ptr %0, i64 312
  store i32 %72, ptr %73, align 8, !tbaa !27
  %74 = load ptr, ptr %20, align 8, !tbaa !16
  %75 = getelementptr inbounds i8, ptr %74, i64 64
  %76 = load ptr, ptr %75, align 8
  %77 = call noundef i64 %76(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 24)
  %78 = trunc i64 %77 to i32
  %79 = getelementptr inbounds i8, ptr %0, i64 316
  store i32 %78, ptr %79, align 4, !tbaa !28
  %80 = getelementptr inbounds i8, ptr %0, i64 320
  %81 = getelementptr inbounds i8, ptr %0, i64 344
  call void @llvm.memset.p0.i64(ptr noundef nonnull align 8 dereferenceable(24) %80, i8 0, i64 24, i1 false)
  store i32 1, ptr %81, align 8, !tbaa !29
  %82 = load ptr, ptr %20, align 8, !tbaa !16
  %83 = getelementptr inbounds i8, ptr %82, i64 64
  %84 = load ptr, ptr %83, align 8
  %85 = call noundef i64 %84(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 0)
  %86 = trunc i64 %85 to i32
  %87 = getelementptr inbounds i8, ptr %0, i64 348
  store i32 %86, ptr %87, align 4, !tbaa !30
  %88 = getelementptr inbounds i8, ptr %0, i64 352
  store i32 0, ptr %88, align 8, !tbaa !31
  %89 = getelementptr inbounds i8, ptr %0, i64 356
  store i32 0, ptr %89, align 4, !tbaa !32
  %90 = getelementptr inbounds i8, ptr %0, i64 360
  store i32 1, ptr %90, align 8, !tbaa !33
  %91 = getelementptr inbounds i8, ptr %0, i64 364
  %92 = getelementptr inbounds i8, ptr %0, i64 416
  store i64 0, ptr %92, align 8, !tbaa !34
  call void @llvm.memset.p0.i64(ptr noundef nonnull align 4 dereferenceable(48) %91, i8 0, i64 48, i1 false)
  %93 = load ptr, ptr %20, align 8, !tbaa !16
  %94 = getelementptr inbounds i8, ptr %93, i64 16
  %95 = load ptr, ptr %94, align 8
  %96 = call noundef i64 %95(ptr noundef nonnull align 8 dereferenceable(8) %20)
  %97 = icmp ugt i64 %96, 1
  %98 = zext i1 %97 to i32
  %99 = getelementptr inbounds i8, ptr %0, i64 424
  store i32 %98, ptr %99, align 8, !tbaa !35
  %100 = getelementptr inbounds i8, ptr %0, i64 428
  call void @llvm.memset.p0.i64(ptr noundef nonnull align 4 dereferenceable(20) %100, i8 0, i64 20, i1 false)
  %101 = load ptr, ptr %20, align 8, !tbaa !16
  %102 = getelementptr inbounds i8, ptr %101, i64 24
  %103 = load ptr, ptr %102, align 8
  %104 = call noundef i64 %103(ptr noundef nonnull align 8 dereferenceable(8) %20)
  %105 = trunc i64 %104 to i32
  %106 = getelementptr inbounds i8, ptr %0, i64 448
  store i32 %105, ptr %106, align 8, !tbaa !36
  %107 = getelementptr inbounds i8, ptr %0, i64 452
  store i32 1, ptr %107, align 4, !tbaa !37
  %108 = getelementptr inbounds i8, ptr %0, i64 456
  store i32 0, ptr %108, align 8, !tbaa !38
  %109 = getelementptr inbounds i8, ptr %0, i64 460
  store i32 0, ptr %109, align 4, !tbaa !39
  %110 = load ptr, ptr %20, align 8, !tbaa !16
  %111 = getelementptr inbounds i8, ptr %110, i64 64
  %112 = load ptr, ptr %111, align 8
  %113 = call noundef i64 %112(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 40)
  %114 = trunc i64 %113 to i32
  %115 = getelementptr inbounds i8, ptr %0, i64 464
  store i32 %114, ptr %115, align 8, !tbaa !40
  %116 = load ptr, ptr %20, align 8, !tbaa !16
  %117 = getelementptr inbounds i8, ptr %116, i64 64
  %118 = load ptr, ptr %117, align 8
  %119 = call noundef i64 %118(ptr noundef nonnull align 8 dereferenceable(8) %20, i32 noundef 8)
  %120 = trunc i64 %119 to i32
  %121 = getelementptr inbounds i8, ptr %0, i64 468
  store i32 %120, ptr %121, align 4, !tbaa !41
  call void @llvm.lifetime.end.p0(i64 1, ptr nonnull %3) #6
  br label %124

122:                                              ; preds = %18
  %123 = landingpad { ptr, i32 }
          cleanup
  call void @_ZNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEED2Ev(ptr noundef nonnull align 8 dereferenceable(32) %4) #6
  call void @llvm.lifetime.end.p0(i64 32, ptr nonnull %4) #6
  call void @llvm.lifetime.end.p0(i64 1, ptr nonnull %3) #6
  resume { ptr, i32 } %123

124:                                              ; preds = %24, %7, %2
  %.1 = phi i32 [ %6, %2 ], [ 0, %24 ], [ 10, %7 ]
  ret i32 %.1
}

declare noundef ptr @_ZNK7hipsycl2rt5pcuda15device_topology10get_deviceEiii(ptr noundef nonnull align 8 dereferenceable(24), i32 noundef, i32 noundef, i32 noundef) local_unnamed_addr #1

; Function Attrs: inlinehint mustprogress uwtable
declare hidden void @"_ZZ24pcudaGetDevicePropertiesENK3$_0clERKNSt7__cxx1112basic_stringIcSt11char_traitsIcESaIcEEEmPc"(ptr noundef nonnull align 1 dereferenceable(1), ptr noundef nonnull align 8 dereferenceable(32), i64 noundef, ptr noundef) local_unnamed_addr #4 align 2

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef nonnull align 8 dereferenceable(8) ptr @_ZNSt6vectorImSaImEEixEm(ptr noundef nonnull align 8 dereferenceable(24), i64 noundef) local_unnamed_addr #2 align 2

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

; Function Attrs: mustprogress nounwind uwtable
declare dso_local noundef i32 @_ZNSt14numeric_limitsIiE3maxEv() local_unnamed_addr #2 align 2

; Function Attrs: nocallback nofree nounwind willreturn memory(argmem: write)
declare void @llvm.memset.p0.i64(ptr nocapture writeonly, i8, i64, i1 immarg) #5

attributes #0 = { mustprogress nocallback nofree nosync nounwind willreturn memory(argmem: readwrite) }
attributes #1 = { "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 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 #3 = { 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 #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: write) }
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, !15, i64 24}
!6 = !{!"_ZTSN7hipsycl2rt5pcuda15device_topology6deviceE", !7, i64 0, !10, i64 4, !15, i64 24}
!7 = !{!"int", !8, i64 0}
!8 = !{!"omnipotent char", !9, i64 0}
!9 = !{!"Simple C++ TBAA"}
!10 = !{!"_ZTSN7hipsycl2rt9device_idE", !11, i64 0, !7, i64 12}
!11 = !{!"_ZTSN7hipsycl2rt18backend_descriptorE", !12, i64 0, !13, i64 4, !14, i64 8}
!12 = !{!"_ZTSN7hipsycl2rt10backend_idE", !8, i64 0}
!13 = !{!"_ZTSN7hipsycl2rt17hardware_platformE", !8, i64 0}
!14 = !{!"_ZTSN7hipsycl2rt12api_platformE", !8, i64 0}
!15 = !{!"any pointer", !8, i64 0}
!16 = !{!17, !17, i64 0}
!17 = !{!"vtable pointer", !9, i64 0}
!18 = !{!19, !20, i64 256}
!19 = !{!"_ZTS15pcudaDeviceProp", !8, i64 0, !20, i64 256, !20, i64 264, !7, i64 272, !7, i64 276, !20, i64 280, !7, i64 288, !8, i64 292, !8, i64 304, !7, i64 316, !20, i64 320, !7, i64 328, !7, i64 332, !20, i64 336, !7, i64 344, !7, i64 348, !7, i64 352, !7, i64 356, !7, i64 360, !7, i64 364, !7, i64 368, !8, i64 372, !8, i64 380, !8, i64 392, !8, i64 400, !20, i64 416, !7, i64 424, !7, i64 428, !7, i64 432, !7, i64 436, !7, i64 440, !7, i64 444, !7, i64 448, !7, i64 452, !7, i64 456, !7, i64 460, !7, i64 464, !7, i64 468}
!20 = !{!"long", !8, i64 0}
!21 = !{!19, !20, i64 264}
!22 = !{!19, !7, i64 272}
!23 = !{!20, !20, i64 0}
!24 = !{!19, !7, i64 276}
!25 = !{!19, !20, i64 280}
!26 = !{!19, !7, i64 288}
!27 = !{!7, !7, i64 0}
!28 = !{!19, !7, i64 316}
!29 = !{!19, !7, i64 344}
!30 = !{!19, !7, i64 348}
!31 = !{!19, !7, i64 352}
!32 = !{!19, !7, i64 356}
!33 = !{!19, !7, i64 360}
!34 = !{!19, !20, i64 416}
!35 = !{!19, !7, i64 424}
!36 = !{!19, !7, i64 448}
!37 = !{!19, !7, i64 452}
!38 = !{!19, !7, i64 456}
!39 = !{!19, !7, i64 460}
!40 = !{!19, !7, i64 464}
!41 = !{!19, !7, i64 468}


; Effective O3 passes: EarlyCSEPass,GlobalOptPass,InferFunctionAttrsPass,InstCombinePass,MemCpyOptPass,PostOrderFunctionAttrsPass,SROAPass,SimplifyCFGPass,TailCallElimPass