Skip to content
Snippets Groups Projects
Commit a5ac493b authored by Prakalp Srivastava's avatar Prakalp Srivastava
Browse files

Preparing test case for GPU

parent c500ad88
No related branches found
No related tags found
No related merge requests found
Hide whitespace changes
Inline Side-by-side
llvm/test/VISC/gemm_opencl/matrixMul/Makefile
+ 1
− 1
View file @ a5ac493b
...@@ -27,4 +27,4 @@ $(HOST:%=%.bin):%.bin:%.c ...@@ -27,4 +27,4 @@ $(HOST:%=%.bin):%.bin:%.c
$(LLVM_CC) -O3 -lOpenCL -I /usr/local/cuda/include $< -o $@ $(LLVM_CC) -O3 -lOpenCL -I /usr/local/cuda/include $< -o $@
clean : clean :
rm -f *.ll *.bc *.s *.bin rm -f $(HOST).ll $(KERNELS).ll *.bc *.s *.bin
llvm/test/VISC/gemm_opencl/matrixMul/gemm_opencl.c
+ 61
− 55
View file @ a5ac493b
...@@ -109,52 +109,9 @@ int checkResults(float* A, float* B, float* C) { ...@@ -109,52 +109,9 @@ int checkResults(float* A, float* B, float* C) {
return 1; // Success return 1; // Success
} }
// Main // GPU Computation of MatrixMul
int main(int argc, char** argv) { void computeMatrixMul(float* h_A, unsigned bytes_A, float* h_B, unsigned bytes_B, float* h_C, unsigned bytes_C) {
// OpenCL specific variables
// seed for rand()
srand(2006);
// Allocate host memory for matrices A and B
unsigned int size_A = WA * HA;
unsigned int bytes_A = sizeof(float) * size_A;
float* h_A = (float*) malloc(bytes_A);
unsigned int size_B = WB * HB;
unsigned int bytes_B = sizeof(float) * size_B;
float* h_B = (float*) malloc(bytes_B);
// Initialize host memory
randomInit(h_A, size_A);
randomInit(h_B, size_B);
/*
// Print A and B
printf("\n\nMatrix A\n");
for(int i = 0; i < size_A; i++)
{
printf("%f ", h_A[i]);
if(((i + 1) % WA) == 0)
printf("\n");
}
printf("\n\nMatrix B\n");
for(int i = 0; i < size_B; i++)
{
printf("%f ", h_B[i]);
if(((i + 1) % WB) == 0)
printf("\n");
}
*/
// Allocate host memory for the result matrix C
unsigned int size_C = WC * HC;
unsigned int bytes_C = sizeof(float) * size_C;
float* h_C = (float*) malloc(bytes_C);
// Initialize OpenCL
// OpenCL specific variables
cl_context clGPUContext; cl_context clGPUContext;
cl_command_queue clCommandQue; cl_command_queue clCommandQue;
cl_program clProgram; cl_program clProgram;
...@@ -291,16 +248,7 @@ int main(int argc, char** argv) { ...@@ -291,16 +248,7 @@ int main(int argc, char** argv) {
} }
printf("\n"); printf("\n");
*/ */
if(checkResults(h_A, h_B, h_C))
printf("\nPass!\n");
else
printf("\nFailed!\n");
printf("\nDone!\n");
// Deallocate memory // Deallocate memory
free(h_A);
free(h_B);
free(h_C);
clReleaseMemObject(d_A); clReleaseMemObject(d_A);
clReleaseMemObject(d_C); clReleaseMemObject(d_C);
...@@ -315,3 +263,61 @@ int main(int argc, char** argv) { ...@@ -315,3 +263,61 @@ int main(int argc, char** argv) {
} }
// Main
int main(int argc, char** argv) {
// seed for rand()
srand(2006);
// Allocate host memory for matrices A and B
unsigned int size_A = WA * HA;
unsigned int bytes_A = sizeof(float) * size_A;
float* h_A = (float*) malloc(bytes_A);
unsigned int size_B = WB * HB;
unsigned int bytes_B = sizeof(float) * size_B;
float* h_B = (float*) malloc(bytes_B);
// Initialize host memory
randomInit(h_A, size_A);
randomInit(h_B, size_B);
/*
// Print A and B
printf("\n\nMatrix A\n");
for(int i = 0; i < size_A; i++)
{
printf("%f ", h_A[i]);
if(((i + 1) % WA) == 0)
printf("\n");
}
printf("\n\nMatrix B\n");
for(int i = 0; i < size_B; i++)
{
printf("%f ", h_B[i]);
if(((i + 1) % WB) == 0)
printf("\n");
}
*/
// Allocate host memory for the result matrix C
unsigned int size_C = WC * HC;
unsigned int bytes_C = sizeof(float) * size_C;
float* h_C = (float*) malloc(bytes_C);
// Compute using OpenCL
computeMatrixMul(h_A, bytes_A, h_B, bytes_B, h_C, bytes_C);
if(checkResults(h_A, h_B, h_C))
printf("\nPass!\n");
else
printf("\nFailed!\n");
printf("\nDone!\n");
// Deallocate memory
free(h_A);
free(h_B);
free(h_C);
}
llvm/test/VISC/gemm_opencl/matrixMul/visc_gemm_opencl.ll 0 → 100644
+ 719
− 0
View file @ a5ac493b
; RUN: opt -load LLVMBuildDFG.so -load LLVMDFG2LLVM_X86.so -load LLVMClearDFG.so -dfg2llvm-x86 -clearDFG -o %t.ll -S < %s
; RUN: llvm-link %t.ll ~/current-src/projects/visc-rt/visc-rt.ll -S -o %t.linked.ll
; RUN: clang -O3 %t.linked.ll -lpthread -lOpenCL -o %t.bin
; RUN: %t.bin 5
; ModuleID = 'gemm_opencl.c'
target datalayout = "e-p:64:64:64-i1:8:8-i8:8:8-i16:16:16-i32:32:32-i64:64:64-f32:32:32-f64:64:64-v64:64:64-v128:128:128-a0:0:64-s0:64:64-f80:128:128-n8:16:32:64-S128"
target triple = "x86_64-unknown-linux-gnu"
%struct._IO_FILE = type { i32, i8*, i8*, i8*, i8*, i8*, i8*, i8*, i8*, i8*, i8*, i8*, %struct._IO_marker*, %struct._IO_FILE*, i32, i32, i64, i16, i8, [1 x i8], i8*, i64, i8*, i8*, i8*, i8*, i64, i32, [20 x i8] }
%struct._IO_marker = type { %struct._IO_marker*, %struct._IO_FILE*, i32 }
%struct._cl_mem = type opaque
%struct._cl_platform_id = type opaque
%struct._cl_context = type opaque
%struct._cl_device_id = type opaque
%struct._cl_command_queue = type opaque
%struct._cl_program = type opaque
%struct._cl_kernel = type opaque
%struct._cl_event = type opaque
; Return Type of VISC Compute Matrix Mul
%rtype = type {float*, i32}
%struct.arg = type { float*, i32, float*, i32, %rtype }
; Function Attrs: nounwind
declare i8* @llvm.visc.launch(i8*, i8*) #0
; Function Attrs: nounwind
declare void @llvm.visc.wait(i8*) #0
@.str = private unnamed_addr constant [3 x i8] c"rb\00", align 1
@.str1 = private unnamed_addr constant [45 x i8] c"Mismatch at %d,%d --- C = %f and goldC = %f\0A\00", align 1
@.str2 = private unnamed_addr constant [35 x i8] c"Failure to get number of platforms\00", align 1
@.str3 = private unnamed_addr constant [28 x i8] c"Failure to get platform IDs\00", align 1
@.str4 = private unnamed_addr constant [12 x i8] c" -- %d --\0A\00", align 1
@.str5 = private unnamed_addr constant [16 x i8] c" PROFILE = %s\0A\00", align 1
@.str6 = private unnamed_addr constant [16 x i8] c" VERSION = %s\0A\00", align 1
@.str7 = private unnamed_addr constant [13 x i8] c" NAME = %s\0A\00", align 1
@.str8 = private unnamed_addr constant [15 x i8] c" VENDOR = %s\0A\00", align 1
@.str9 = private unnamed_addr constant [19 x i8] c" EXTENSIONS = %s\0A\00", align 1
@.str10 = private unnamed_addr constant [30 x i8] c"Failure to create GPU context\00", align 1
@.str11 = private unnamed_addr constant [28 x i8] c"Failure to get context info\00", align 1
@.str12 = private unnamed_addr constant [32 x i8] c"Failure to create command queue\00", align 1
@.str13 = private unnamed_addr constant [18 x i8] c"matrixMul.nvptx.s\00", align 1
@.str14 = private unnamed_addr constant [31 x i8] c"Failure to load Program Binary\00", align 1
@.str15 = private unnamed_addr constant [38 x i8] c"Failure to create program from binary\00", align 1
@.str16 = private unnamed_addr constant [25 x i8] c"Failure to build program\00", align 1
@.str17 = private unnamed_addr constant [10 x i8] c"matrixMul\00", align 1
@.str18 = private unnamed_addr constant [25 x i8] c"Failure to create kernel\00", align 1
@.str19 = private unnamed_addr constant [32 x i8] c"Failure to set kernel arguments\00", align 1
@.str20 = private unnamed_addr constant [26 x i8] c"Failure to enqueue kernel\00", align 1
@.str21 = private unnamed_addr constant [23 x i8] c"Failure to read buffer\00", align 1
@stderr = external global %struct._IO_FILE*
@.str25 = private unnamed_addr constant [11 x i8] c"ERROR: %s\0A\00", align 1
@str = private unnamed_addr constant [9 x i8] c"\0AFailed!\00"
@str26 = private unnamed_addr constant [7 x i8] c"\0ADone!\00"
@str27 = private unnamed_addr constant [7 x i8] c"\0APass!\00"
; Function Attrs: nounwind uwtable
define void @randomInit(float* nocapture %data, i32 %size) #0 {
entry:
%cmp3 = icmp sgt i32 %size, 0
br i1 %cmp3, label %for.body, label %for.end
for.body: ; preds = %entry, %for.body
%indvars.iv = phi i64 [ %indvars.iv.next, %for.body ], [ 0, %entry ]
%call = tail call i32 @rand() #4
%conv = sitofp i32 %call to float
%div = fmul float %conv, 0x3E00000000000000
%arrayidx = getelementptr inbounds float* %data, i64 %indvars.iv
store float %div, float* %arrayidx, align 4, !tbaa !0
%indvars.iv.next = add i64 %indvars.iv, 1
%lftr.wideiv = trunc i64 %indvars.iv.next to i32
%exitcond = icmp eq i32 %lftr.wideiv, %size
br i1 %exitcond, label %for.end, label %for.body
for.end: ; preds = %for.body, %entry
ret void
}
; Function Attrs: nounwind
declare i32 @rand() #1
; Function Attrs: nounwind uwtable
define noalias i8* @LoadProgSource(i8* nocapture %cFilename, i64* %szFinalLength) #0 {
entry:
%call = tail call %struct._IO_FILE* @fopen(i8* %cFilename, i8* getelementptr inbounds ([3 x i8]* @.str, i64 0, i64 0)) #4
%cmp = icmp eq %struct._IO_FILE* %call, null
br i1 %cmp, label %return, label %if.end
if.end: ; preds = %entry
%call1 = tail call i32 @fseek(%struct._IO_FILE* %call, i64 0, i32 2) #4
%call2 = tail call i64 @ftell(%struct._IO_FILE* %call) #4
%call3 = tail call i32 @fseek(%struct._IO_FILE* %call, i64 0, i32 0) #4
%add = add i64 %call2, 1
%call4 = tail call noalias i8* @malloc(i64 %add) #4
%call5 = tail call i64 @fread(i8* %call4, i64 %call2, i64 1, %struct._IO_FILE* %call) #4
%cmp6 = icmp eq i64 %call5, 1
%call8 = tail call i32 @fclose(%struct._IO_FILE* %call) #4
br i1 %cmp6, label %if.end9, label %if.then7
if.then7: ; preds = %if.end
tail call void @free(i8* %call4) #4
br label %return
if.end9: ; preds = %if.end
%cmp11 = icmp eq i64* %szFinalLength, null
br i1 %cmp11, label %if.end13, label %if.then12
if.then12: ; preds = %if.end9
store i64 %call2, i64* %szFinalLength, align 8, !tbaa !3
br label %if.end13
if.end13: ; preds = %if.end9, %if.then12
%arrayidx = getelementptr inbounds i8* %call4, i64 %call2
store i8 0, i8* %arrayidx, align 1, !tbaa !1
br label %return
return: ; preds = %entry, %if.end13, %if.then7
%retval.0 = phi i8* [ null, %if.then7 ], [ %call4, %if.end13 ], [ null, %entry ]
ret i8* %retval.0
}
; Function Attrs: nounwind
declare noalias %struct._IO_FILE* @fopen(i8* nocapture, i8* nocapture) #1
; Function Attrs: nounwind
declare i32 @fseek(%struct._IO_FILE* nocapture, i64, i32) #1
; Function Attrs: nounwind
declare i64 @ftell(%struct._IO_FILE* nocapture) #1
; Function Attrs: nounwind
declare noalias i8* @malloc(i64) #1
; Function Attrs: nounwind
declare i64 @fread(i8* nocapture, i64, i64, %struct._IO_FILE* nocapture) #1
; Function Attrs: nounwind
declare i32 @fclose(%struct._IO_FILE* nocapture) #1
; Function Attrs: nounwind
declare void @free(i8* nocapture) #1
; Function Attrs: nounwind readnone uwtable
define i32 @isEqual(float %a, float %b) #2 {
entry:
%sub = fsub float %a, %b
%fabsf = tail call float @fabsf(float %sub) #6
%0 = fpext float %fabsf to double
%cmp = fcmp olt double %0, 1.000000e-03
%conv1 = zext i1 %cmp to i32
ret i32 %conv1
}
; Function Attrs: nounwind uwtable
define i32 @checkResults(float* nocapture %A, float* nocapture %B, float* nocapture %C) #0 {
entry:
br label %for.cond4.preheader
for.cond4.preheader: ; preds = %entry, %for.inc50
%indvars.iv92 = phi i64 [ 0, %entry ], [ %indvars.iv.next93, %for.inc50 ]
%i.081 = phi i32 [ 0, %entry ], [ %inc51, %for.inc50 ]
%0 = shl nsw i64 %indvars.iv92, 10
br label %for.body7
for.cond4: ; preds = %for.end
%inc48 = add nsw i32 %j.079, 1
%1 = trunc i64 %indvars.iv.next89 to i32
%cmp5 = icmp slt i32 %1, 1024
br i1 %cmp5, label %for.body7, label %for.inc50
for.body7: ; preds = %for.cond4.preheader, %for.cond4
%indvars.iv88 = phi i64 [ 0, %for.cond4.preheader ], [ %indvars.iv.next89, %for.cond4 ]
%j.079 = phi i32 [ 0, %for.cond4.preheader ], [ %inc48, %for.cond4 ]
%2 = add nsw i64 %indvars.iv88, %0
br label %for.body12
for.body12: ; preds = %for.body12, %for.body7
%indvars.iv = phi i64 [ 0, %for.body7 ], [ %indvars.iv.next, %for.body12 ]
%3 = phi float [ 0.000000e+00, %for.body7 ], [ %add26, %for.body12 ]
%4 = add nsw i64 %indvars.iv, %0
%arrayidx16 = getelementptr inbounds float* %A, i64 %4
%5 = load float* %arrayidx16, align 4, !tbaa !0
%6 = shl i64 %indvars.iv, 10
%7 = add nsw i64 %6, %indvars.iv88
%arrayidx20 = getelementptr inbounds float* %B, i64 %7
%8 = load float* %arrayidx20, align 4, !tbaa !0
%mul21 = fmul float %5, %8
%add26 = fadd float %3, %mul21
%indvars.iv.next = add i64 %indvars.iv, 1
%lftr.wideiv = trunc i64 %indvars.iv.next to i32
%exitcond = icmp eq i32 %lftr.wideiv, 1024
br i1 %exitcond, label %for.end, label %for.body12
for.end: ; preds = %for.body12
%arrayidx34 = getelementptr inbounds float* %C, i64 %2
%9 = load float* %arrayidx34, align 4, !tbaa !0
%sub.i = fsub float %add26, %9
%fabsf.i = tail call float @fabsf(float %sub.i) #6
%10 = fpext float %fabsf.i to double
%cmp.i = fcmp olt double %10, 1.000000e-03
%indvars.iv.next89 = add i64 %indvars.iv88, 1
br i1 %cmp.i, label %for.cond4, label %if.then
if.then: ; preds = %for.end
%conv40 = fpext float %9 to double
%conv45 = fpext float %add26 to double
%call46 = tail call i32 (i8*, ...)* @printf(i8* getelementptr inbounds ([45 x i8]* @.str1, i64 0, i64 0), i32 %i.081, i32 %j.079, double %conv40, double %conv45) #4
br label %return
for.inc50: ; preds = %for.cond4
%indvars.iv.next93 = add i64 %indvars.iv92, 1
%inc51 = add nsw i32 %i.081, 1
%11 = trunc i64 %indvars.iv.next93 to i32
%cmp = icmp slt i32 %11, 1024
br i1 %cmp, label %for.cond4.preheader, label %return
return: ; preds = %for.inc50, %if.then
%retval.0 = phi i32 [ 0, %if.then ], [ 1, %for.inc50 ]
ret i32 %retval.0
}
; Function Attrs: nounwind
declare i32 @printf(i8* nocapture, ...) #1
; Function Attrs: nounwind uwtable
define void @computeMatrixMul(float* %h_A, i32 %bytes_A, float* %h_B, i32 %bytes_B, float* %h_C, i32 %bytes_C) #0 {
entry:
%dataBytes = alloca i64, align 8
%errcode = alloca i32, align 4
%d_A = alloca %struct._cl_mem*, align 8
%d_B = alloca %struct._cl_mem*, align 8
%d_C = alloca %struct._cl_mem*, align 8
%numPlatforms = alloca i32, align 4
%buffer = alloca [10240 x i8], align 16
%properties = alloca [3 x i64], align 16
%binaryLength = alloca i64, align 8
%clMatrixMul = alloca i8*, align 8
%binaryStatus = alloca i32, align 4
%localWorkSize = alloca [2 x i64], align 16
%globalWorkSize = alloca [2 x i64], align 16
%wA = alloca i32, align 4
%wC = alloca i32, align 4
%call = call i32 @clGetPlatformIDs(i32 0, %struct._cl_platform_id** null, i32* %numPlatforms) #4
store i32 %call, i32* %errcode, align 4, !tbaa !4
%cmp.i = icmp eq i32 %call, 0
br i1 %cmp.i, label %checkErr.exit, label %if.then.i
if.then.i: ; preds = %entry
%0 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %0, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([35 x i8]* @.str2, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit: ; preds = %entry
%1 = load i32* %numPlatforms, align 4, !tbaa !4
%2 = zext i32 %1 to i64
%vla = alloca %struct._cl_platform_id*, i64 %2, align 16
%call1 = call i32 @clGetPlatformIDs(i32 %1, %struct._cl_platform_id** %vla, i32* null) #4
store i32 %call1, i32* %errcode, align 4, !tbaa !4
%cmp.i105 = icmp eq i32 %call1, 0
br i1 %cmp.i105, label %for.cond.preheader, label %if.then.i107
for.cond.preheader: ; preds = %checkErr.exit
%3 = load i32* %numPlatforms, align 4, !tbaa !4
%cmp148 = icmp eq i32 %3, 0
br i1 %cmp148, label %for.end, label %for.body.lr.ph
for.body.lr.ph: ; preds = %for.cond.preheader
%4 = getelementptr inbounds [10240 x i8]* %buffer, i64 0, i64 0
br label %for.body
if.then.i107: ; preds = %checkErr.exit
%5 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i106 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %5, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([28 x i8]* @.str3, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
for.body: ; preds = %for.body.lr.ph, %for.body
%i.0149 = phi i32 [ 0, %for.body.lr.ph ], [ %inc, %for.body ]
call void @llvm.lifetime.start(i64 10240, i8* %4) #4
%call2 = call i32 (i8*, ...)* @printf(i8* getelementptr inbounds ([12 x i8]* @.str4, i64 0, i64 0), i32 %i.0149) #4
%idxprom = zext i32 %i.0149 to i64
%arrayidx = getelementptr inbounds %struct._cl_platform_id** %vla, i64 %idxprom
%6 = load %struct._cl_platform_id** %arrayidx, align 8, !tbaa !5
%call3 = call i32 @clGetPlatformInfo(%struct._cl_platform_id* %6, i32 2304, i64 10240, i8* %4, i64* null) #4
%call5 = call i32 (i8*, ...)* @printf(i8* getelementptr inbounds ([16 x i8]* @.str5, i64 0, i64 0), i8* %4) #4
%7 = load %struct._cl_platform_id** %arrayidx, align 8, !tbaa !5
%call9 = call i32 @clGetPlatformInfo(%struct._cl_platform_id* %7, i32 2305, i64 10240, i8* %4, i64* null) #4
%call11 = call i32 (i8*, ...)* @printf(i8* getelementptr inbounds ([16 x i8]* @.str6, i64 0, i64 0), i8* %4) #4
%8 = load %struct._cl_platform_id** %arrayidx, align 8, !tbaa !5
%call15 = call i32 @clGetPlatformInfo(%struct._cl_platform_id* %8, i32 2306, i64 10240, i8* %4, i64* null) #4
%call17 = call i32 (i8*, ...)* @printf(i8* getelementptr inbounds ([13 x i8]* @.str7, i64 0, i64 0), i8* %4) #4
%9 = load %struct._cl_platform_id** %arrayidx, align 8, !tbaa !5
%call21 = call i32 @clGetPlatformInfo(%struct._cl_platform_id* %9, i32 2307, i64 10240, i8* %4, i64* null) #4
%call23 = call i32 (i8*, ...)* @printf(i8* getelementptr inbounds ([15 x i8]* @.str8, i64 0, i64 0), i8* %4) #4
%10 = load %struct._cl_platform_id** %arrayidx, align 8, !tbaa !5
%call27 = call i32 @clGetPlatformInfo(%struct._cl_platform_id* %10, i32 2308, i64 10240, i8* %4, i64* null) #4
%call29 = call i32 (i8*, ...)* @printf(i8* getelementptr inbounds ([19 x i8]* @.str9, i64 0, i64 0), i8* %4) #4
call void @llvm.lifetime.end(i64 10240, i8* %4) #4
%inc = add i32 %i.0149, 1
%11 = load i32* %numPlatforms, align 4, !tbaa !4
%cmp = icmp ult i32 %inc, %11
br i1 %cmp, label %for.body, label %for.end
for.end: ; preds = %for.body, %for.cond.preheader
%arrayinit.begin = getelementptr inbounds [3 x i64]* %properties, i64 0, i64 0
store i64 4228, i64* %arrayinit.begin, align 16, !tbaa !3
%arrayinit.element = getelementptr inbounds [3 x i64]* %properties, i64 0, i64 1
%12 = load %struct._cl_platform_id** %vla, align 16, !tbaa !5
%13 = ptrtoint %struct._cl_platform_id* %12 to i64
%sext = shl i64 %13, 32
%conv = ashr exact i64 %sext, 32
store i64 %conv, i64* %arrayinit.element, align 8, !tbaa !3
%arrayinit.element31 = getelementptr inbounds [3 x i64]* %properties, i64 0, i64 2
store i64 0, i64* %arrayinit.element31, align 16, !tbaa !3
%call33 = call %struct._cl_context* @clCreateContextFromType(i64* %arrayinit.begin, i64 4, void (i8*, i8*, i64, i8*)* null, i8* null, i32* %errcode) #4
%14 = load i32* %errcode, align 4, !tbaa !4
%cmp.i109 = icmp eq i32 %14, 0
br i1 %cmp.i109, label %checkErr.exit112, label %if.then.i111
if.then.i111: ; preds = %for.end
%15 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i110 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %15, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([30 x i8]* @.str10, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit112: ; preds = %for.end
%call34 = call i32 @clGetContextInfo(%struct._cl_context* %call33, i32 4225, i64 0, i8* null, i64* %dataBytes) #4
store i32 %call34, i32* %errcode, align 4, !tbaa !4
%16 = load i64* %dataBytes, align 8, !tbaa !3
%call35 = call noalias i8* @malloc(i64 %16) #4
%17 = bitcast i8* %call35 to %struct._cl_device_id**
%call36 = call i32 @clGetContextInfo(%struct._cl_context* %call33, i32 4225, i64 %16, i8* %call35, i64* null) #4
%18 = load i32* %errcode, align 4, !tbaa !4
%or = or i32 %18, %call36
store i32 %or, i32* %errcode, align 4, !tbaa !4
%cmp.i113 = icmp eq i32 %or, 0
br i1 %cmp.i113, label %checkErr.exit116, label %if.then.i115
if.then.i115: ; preds = %checkErr.exit112
%19 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i114 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %19, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([28 x i8]* @.str11, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit116: ; preds = %checkErr.exit112
%20 = load %struct._cl_device_id** %17, align 8, !tbaa !5
%call38 = call %struct._cl_command_queue* @clCreateCommandQueue(%struct._cl_context* %call33, %struct._cl_device_id* %20, i64 0, i32* %errcode) #4
%21 = load i32* %errcode, align 4, !tbaa !4
%cmp.i117 = icmp eq i32 %21, 0
br i1 %cmp.i117, label %checkErr.exit120, label %if.then.i119
if.then.i119: ; preds = %checkErr.exit116
%22 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i118 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %22, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([32 x i8]* @.str12, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit120: ; preds = %checkErr.exit116
%conv39 = zext i32 %bytes_C to i64
%call40 = call %struct._cl_mem* @clCreateBuffer(%struct._cl_context* %call33, i64 1, i64 %conv39, i8* null, i32* %errcode) #4
store %struct._cl_mem* %call40, %struct._cl_mem** %d_C, align 8, !tbaa !5
%conv41 = zext i32 %bytes_A to i64
%23 = bitcast float* %h_A to i8*
%call42 = call %struct._cl_mem* @clCreateBuffer(%struct._cl_context* %call33, i64 33, i64 %conv41, i8* %23, i32* %errcode) #4
store %struct._cl_mem* %call42, %struct._cl_mem** %d_A, align 8, !tbaa !5
%conv43 = zext i32 %bytes_B to i64
%24 = bitcast float* %h_B to i8*
%call44 = call %struct._cl_mem* @clCreateBuffer(%struct._cl_context* %call33, i64 33, i64 %conv43, i8* %24, i32* %errcode) #4
store %struct._cl_mem* %call44, %struct._cl_mem** %d_B, align 8, !tbaa !5
%call45 = call i8* @LoadProgSource(i8* getelementptr inbounds ([18 x i8]* @.str13, i64 0, i64 0), i64* %binaryLength)
store i8* %call45, i8** %clMatrixMul, align 8, !tbaa !5
%cmp46 = icmp eq i8* %call45, null
br i1 %cmp46, label %if.then.i122, label %checkErr.exit123
if.then.i122: ; preds = %checkErr.exit120
%25 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i121 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %25, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([31 x i8]* @.str14, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit123: ; preds = %checkErr.exit120
%call49 = call %struct._cl_program* @clCreateProgramWithBinary(%struct._cl_context* %call33, i32 1, %struct._cl_device_id** %17, i64* %binaryLength, i8** %clMatrixMul, i32* %binaryStatus, i32* %errcode) #4
%26 = load i32* %errcode, align 4, !tbaa !4
%cmp.i124 = icmp eq i32 %26, 0
br i1 %cmp.i124, label %checkErr.exit127, label %if.then.i126
if.then.i126: ; preds = %checkErr.exit123
%27 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i125 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %27, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([38 x i8]* @.str15, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit127: ; preds = %checkErr.exit123
%call50 = call i32 @clBuildProgram(%struct._cl_program* %call49, i32 0, %struct._cl_device_id** null, i8* null, void (%struct._cl_program*, i8*)* null, i8* null) #4
store i32 %call50, i32* %errcode, align 4, !tbaa !4
%cmp.i128 = icmp eq i32 %call50, 0
br i1 %cmp.i128, label %checkErr.exit131, label %if.then.i130
if.then.i130: ; preds = %checkErr.exit127
%28 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i129 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %28, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([25 x i8]* @.str16, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit131: ; preds = %checkErr.exit127
%call51 = call %struct._cl_kernel* @clCreateKernel(%struct._cl_program* %call49, i8* getelementptr inbounds ([10 x i8]* @.str17, i64 0, i64 0), i32* %errcode) #4
%29 = load i32* %errcode, align 4, !tbaa !4
%cmp.i132 = icmp eq i32 %29, 0
br i1 %cmp.i132, label %checkErr.exit135, label %if.then.i134
if.then.i134: ; preds = %checkErr.exit131
%30 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i133 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %30, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([25 x i8]* @.str18, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit135: ; preds = %checkErr.exit131
store i32 1024, i32* %wA, align 4, !tbaa !4
store i32 1024, i32* %wC, align 4, !tbaa !4
%31 = bitcast %struct._cl_mem** %d_C to i8*
%call52 = call i32 @clSetKernelArg(%struct._cl_kernel* %call51, i32 0, i64 8, i8* %31) #4
store i32 %call52, i32* %errcode, align 4, !tbaa !4
%32 = bitcast %struct._cl_mem** %d_A to i8*
%call53 = call i32 @clSetKernelArg(%struct._cl_kernel* %call51, i32 1, i64 8, i8* %32) #4
%33 = load i32* %errcode, align 4, !tbaa !4
%or54 = or i32 %33, %call53
store i32 %or54, i32* %errcode, align 4, !tbaa !4
%34 = bitcast %struct._cl_mem** %d_B to i8*
%call55 = call i32 @clSetKernelArg(%struct._cl_kernel* %call51, i32 2, i64 8, i8* %34) #4
%35 = load i32* %errcode, align 4, !tbaa !4
%or56 = or i32 %35, %call55
store i32 %or56, i32* %errcode, align 4, !tbaa !4
%36 = bitcast i32* %wA to i8*
%call57 = call i32 @clSetKernelArg(%struct._cl_kernel* %call51, i32 3, i64 4, i8* %36) #4
%37 = load i32* %errcode, align 4, !tbaa !4
%or58 = or i32 %37, %call57
store i32 %or58, i32* %errcode, align 4, !tbaa !4
%38 = bitcast i32* %wC to i8*
%call59 = call i32 @clSetKernelArg(%struct._cl_kernel* %call51, i32 4, i64 4, i8* %38) #4
%39 = load i32* %errcode, align 4, !tbaa !4
%or60 = or i32 %39, %call59
store i32 %or60, i32* %errcode, align 4, !tbaa !4
%cmp.i136 = icmp eq i32 %or60, 0
br i1 %cmp.i136, label %checkErr.exit139, label %if.then.i138
if.then.i138: ; preds = %checkErr.exit135
%40 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i137 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %40, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([32 x i8]* @.str19, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit139: ; preds = %checkErr.exit135
%arrayidx61 = getelementptr inbounds [2 x i64]* %localWorkSize, i64 0, i64 0
store i64 16, i64* %arrayidx61, align 16, !tbaa !3
%arrayidx62 = getelementptr inbounds [2 x i64]* %localWorkSize, i64 0, i64 1
store i64 16, i64* %arrayidx62, align 8, !tbaa !3
%arrayidx63 = getelementptr inbounds [2 x i64]* %globalWorkSize, i64 0, i64 0
store i64 1024, i64* %arrayidx63, align 16, !tbaa !3
%arrayidx64 = getelementptr inbounds [2 x i64]* %globalWorkSize, i64 0, i64 1
store i64 1024, i64* %arrayidx64, align 8, !tbaa !3
%call67 = call i32 @clEnqueueNDRangeKernel(%struct._cl_command_queue* %call38, %struct._cl_kernel* %call51, i32 2, i64* null, i64* %arrayidx63, i64* %arrayidx61, i32 0, %struct._cl_event** null, %struct._cl_event** null) #4
store i32 %call67, i32* %errcode, align 4, !tbaa !4
%cmp.i140 = icmp eq i32 %call67, 0
br i1 %cmp.i140, label %checkErr.exit143, label %if.then.i142
if.then.i142: ; preds = %checkErr.exit139
%41 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i141 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %41, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([26 x i8]* @.str20, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit143: ; preds = %checkErr.exit139
%42 = load %struct._cl_mem** %d_C, align 8, !tbaa !5
%43 = bitcast float* %h_C to i8*
%call69 = call i32 @clEnqueueReadBuffer(%struct._cl_command_queue* %call38, %struct._cl_mem* %42, i32 1, i64 0, i64 %conv39, i8* %43, i32 0, %struct._cl_event** null, %struct._cl_event** null) #4
store i32 %call69, i32* %errcode, align 4, !tbaa !4
%cmp.i144 = icmp eq i32 %call69, 0
br i1 %cmp.i144, label %checkErr.exit147, label %if.then.i146
if.then.i146: ; preds = %checkErr.exit143
%44 = load %struct._IO_FILE** @stderr, align 8, !tbaa !5
%call.i145 = call i32 (%struct._IO_FILE*, i8*, ...)* @fprintf(%struct._IO_FILE* %44, i8* getelementptr inbounds ([11 x i8]* @.str25, i64 0, i64 0), i8* getelementptr inbounds ([23 x i8]* @.str21, i64 0, i64 0)) #4
call void @exit(i32 1) #7
unreachable
checkErr.exit147: ; preds = %checkErr.exit143
%45 = load %struct._cl_mem** %d_A, align 8, !tbaa !5
%call70 = call i32 @clReleaseMemObject(%struct._cl_mem* %45) #4
%46 = load %struct._cl_mem** %d_C, align 8, !tbaa !5
%call71 = call i32 @clReleaseMemObject(%struct._cl_mem* %46) #4
%47 = load %struct._cl_mem** %d_B, align 8, !tbaa !5
%call72 = call i32 @clReleaseMemObject(%struct._cl_mem* %47) #4
call void @free(i8* %call35) #4
%48 = load i8** %clMatrixMul, align 8, !tbaa !5
call void @free(i8* %48) #4
%call73 = call i32 @clReleaseContext(%struct._cl_context* %call33) #4
%call74 = call i32 @clReleaseKernel(%struct._cl_kernel* %call51) #4
%call75 = call i32 @clReleaseProgram(%struct._cl_program* %call49) #4
%call76 = call i32 @clReleaseCommandQueue(%struct._cl_command_queue* %call38) #4
ret void
}
declare i32 @clGetPlatformIDs(i32, %struct._cl_platform_id**, i32*) #3
; Function Attrs: nounwind
declare void @llvm.lifetime.start(i64, i8* nocapture) #4
declare i32 @clGetPlatformInfo(%struct._cl_platform_id*, i32, i64, i8*, i64*) #3
; Function Attrs: nounwind
declare void @llvm.lifetime.end(i64, i8* nocapture) #4
declare %struct._cl_context* @clCreateContextFromType(i64*, i64, void (i8*, i8*, i64, i8*)*, i8*, i32*) #3
declare i32 @clGetContextInfo(%struct._cl_context*, i32, i64, i8*, i64*) #3
declare %struct._cl_command_queue* @clCreateCommandQueue(%struct._cl_context*, %struct._cl_device_id*, i64, i32*) #3
declare %struct._cl_mem* @clCreateBuffer(%struct._cl_context*, i64, i64, i8*, i32*) #3
declare %struct._cl_program* @clCreateProgramWithBinary(%struct._cl_context*, i32, %struct._cl_device_id**, i64*, i8**, i32*, i32*) #3
declare i32 @clBuildProgram(%struct._cl_program*, i32, %struct._cl_device_id**, i8*, void (%struct._cl_program*, i8*)*, i8*) #3
declare %struct._cl_kernel* @clCreateKernel(%struct._cl_program*, i8*, i32*) #3
declare i32 @clSetKernelArg(%struct._cl_kernel*, i32, i64, i8*) #3
declare i32 @clEnqueueNDRangeKernel(%struct._cl_command_queue*, %struct._cl_kernel*, i32, i64*, i64*, i64*, i32, %struct._cl_event**, %struct._cl_event**) #3
declare i32 @clEnqueueReadBuffer(%struct._cl_command_queue*, %struct._cl_mem*, i32, i64, i64, i8*, i32, %struct._cl_event**, %struct._cl_event**) #3
declare i32 @clReleaseMemObject(%struct._cl_mem*) #3
declare i32 @clReleaseContext(%struct._cl_context*) #3
declare i32 @clReleaseKernel(%struct._cl_kernel*) #3
declare i32 @clReleaseProgram(%struct._cl_program*) #3
declare i32 @clReleaseCommandQueue(%struct._cl_command_queue*) #3
; Function Attrs: nounwind uwtable
define i32 @main(i32 %argc, i8** nocapture %argv) #0 {
entry:
tail call void @srand(i32 2006) #4
%call = tail call noalias i8* @malloc(i64 4194304) #4
%0 = bitcast i8* %call to float*
%call7 = tail call noalias i8* @malloc(i64 4194304) #4
br label %for.body.i
for.body.i: ; preds = %for.body.i, %entry
%indvars.iv.i = phi i64 [ %indvars.iv.next.i, %for.body.i ], [ 0, %entry ]
%call.i = tail call i32 @rand() #4
%conv.i = sitofp i32 %call.i to float
%div.i = fmul float %conv.i, 0x3E00000000000000
%arrayidx.i = getelementptr inbounds float* %0, i64 %indvars.iv.i
store float %div.i, float* %arrayidx.i, align 4, !tbaa !0
%indvars.iv.next.i = add i64 %indvars.iv.i, 1
%lftr.wideiv59 = trunc i64 %indvars.iv.next.i to i32
%exitcond60 = icmp eq i32 %lftr.wideiv59, 1048576
br i1 %exitcond60, label %for.body.i40.preheader, label %for.body.i
for.body.i40.preheader: ; preds = %for.body.i
%1 = bitcast i8* %call7 to float*
br label %for.body.i40
for.body.i40: ; preds = %for.body.i40.preheader, %for.body.i40
%indvars.iv.i32 = phi i64 [ %indvars.iv.next.i37, %for.body.i40 ], [ 0, %for.body.i40.preheader ]
%call.i33 = tail call i32 @rand() #4
%conv.i34 = sitofp i32 %call.i33 to float
%div.i35 = fmul float %conv.i34, 0x3E00000000000000
%arrayidx.i36 = getelementptr inbounds float* %1, i64 %indvars.iv.i32
store float %div.i35, float* %arrayidx.i36, align 4, !tbaa !0
%indvars.iv.next.i37 = add i64 %indvars.iv.i32, 1
%lftr.wideiv57 = trunc i64 %indvars.iv.next.i37 to i32
%exitcond58 = icmp eq i32 %lftr.wideiv57, 1048576
br i1 %exitcond58, label %randomInit.exit41, label %for.body.i40
randomInit.exit41: ; preds = %for.body.i40
%call12 = tail call noalias i8* @malloc(i64 4194304) #4
%2 = bitcast i8* %call12 to float*
; ----- Adding VISC Launch Call -----
; Setting up launch input args
%in.addr = alloca %struct.arg
; Store arguments
%in.addr.h_A = getelementptr %struct.arg* %in.addr, i32 0, i32 0
%in.addr.bytes_A = getelementptr %struct.arg* %in.addr, i32 0, i32 1
%in.addr.h_B = getelementptr %struct.arg* %in.addr, i32 0, i32 2
%in.addr.bytes_B = getelementptr %struct.arg* %in.addr, i32 0, i32 3
store float* %0, float** %in.addr.h_A
store i32 4194304, i32* %in.addr.bytes_A
store float* %1, float** %in.addr.h_B
store i32 4194304, i32* %in.addr.bytes_B
; Change type to i8* and VISC Launch call
%args = bitcast { float*, i32, float*, i32, %rtype }* %in.addr to i8*
%graphID = call i8* @llvm.visc.launch(i8* bitcast (%rtype (float*, i32, float*, i32)* @MatrixMulRoot to i8*), i8* %args)
;tail call void @computeMatrixMul(float* %0, i32 4194304, float* %1, i32 4194304, float* %2, i32 4194304)
; Wait for result
call void @llvm.visc.wait(i8* %graphID)
; Get the result
%out.addr = getelementptr %struct.arg* %in.addr, i32 0, i32 4
%out = load %rtype* %out.addr
%out.h_C = extractvalue %rtype %out, 0
;%2 = extractvalue %rtype %out, 0
%out.bytes_C = extractvalue %rtype %outputstruct, 1
; ----- Completed VISC Launch Call Code -----
br label %for.cond4.preheader.i
for.cond4.preheader.i: ; preds = %for.inc50.i, %randomInit.exit41
%indvars.iv92.i = phi i64 [ 0, %randomInit.exit41 ], [ %indvars.iv.next93.i, %for.inc50.i ]
%i.081.i = phi i32 [ 0, %randomInit.exit41 ], [ %inc51.i, %for.inc50.i ]
%3 = shl nsw i64 %indvars.iv92.i, 10
br label %for.body7.i
for.cond4.i: ; preds = %for.end.i
%inc48.i = add nsw i32 %j.079.i, 1
%4 = trunc i64 %indvars.iv.next89.i to i32
%cmp5.i = icmp slt i32 %4, 1024
br i1 %cmp5.i, label %for.body7.i, label %for.inc50.i
for.body7.i: ; preds = %for.cond4.i, %for.cond4.preheader.i
%indvars.iv88.i = phi i64 [ 0, %for.cond4.preheader.i ], [ %indvars.iv.next89.i, %for.cond4.i ]
%j.079.i = phi i32 [ 0, %for.cond4.preheader.i ], [ %inc48.i, %for.cond4.i ]
br label %for.body12.i
for.body12.i: ; preds = %for.body12.i, %for.body7.i
%indvars.iv.i42 = phi i64 [ 0, %for.body7.i ], [ %indvars.iv.next.i43, %for.body12.i ]
%5 = phi float [ 0.000000e+00, %for.body7.i ], [ %add26.i, %for.body12.i ]
%6 = add nsw i64 %indvars.iv.i42, %3
%arrayidx16.i = getelementptr inbounds float* %0, i64 %6
%7 = load float* %arrayidx16.i, align 4, !tbaa !0
%8 = shl i64 %indvars.iv.i42, 10
%9 = add nsw i64 %8, %indvars.iv88.i
%arrayidx20.i = getelementptr inbounds float* %1, i64 %9
%10 = load float* %arrayidx20.i, align 4, !tbaa !0
%mul21.i = fmul float %7, %10
%add26.i = fadd float %5, %mul21.i
%indvars.iv.next.i43 = add i64 %indvars.iv.i42, 1
%lftr.wideiv = trunc i64 %indvars.iv.next.i43 to i32
%exitcond = icmp eq i32 %lftr.wideiv, 1024
br i1 %exitcond, label %for.end.i, label %for.body12.i
for.end.i: ; preds = %for.body12.i
%11 = add nsw i64 %indvars.iv88.i, %3
;%arrayidx34.i = getelementptr inbounds float* %2, i64 %11
%arrayidx34.i = getelementptr inbounds float* %out.h_C, i64 %11
%12 = load float* %arrayidx34.i, align 4, !tbaa !0
%sub.i.i = fsub float %add26.i, %12
%fabsf.i.i = tail call float @fabsf(float %sub.i.i) #6
%13 = fpext float %fabsf.i.i to double
%cmp.i.i = fcmp olt double %13, 1.000000e-03
%indvars.iv.next89.i = add i64 %indvars.iv88.i, 1
br i1 %cmp.i.i, label %for.cond4.i, label %if.else
for.inc50.i: ; preds = %for.cond4.i
%indvars.iv.next93.i = add i64 %indvars.iv92.i, 1
%inc51.i = add nsw i32 %i.081.i, 1
%14 = trunc i64 %indvars.iv.next93.i to i32
%cmp.i = icmp slt i32 %14, 1024
br i1 %cmp.i, label %for.cond4.preheader.i, label %if.then
if.then: ; preds = %for.inc50.i
%puts31 = tail call i32 @puts(i8* getelementptr inbounds ([7 x i8]* @str27, i64 0, i64 0))
br label %if.end
if.else: ; preds = %for.end.i
%conv40.i = fpext float %12 to double
%conv45.i = fpext float %add26.i to double
%call46.i = tail call i32 (i8*, ...)* @printf(i8* getelementptr inbounds ([45 x i8]* @.str1, i64 0, i64 0), i32 %i.081.i, i32 %j.079.i, double %conv40.i, double %conv45.i) #4
%puts = tail call i32 @puts(i8* getelementptr inbounds ([9 x i8]* @str, i64 0, i64 0))
br label %if.end
if.end: ; preds = %if.else, %if.then
%puts30 = tail call i32 @puts(i8* getelementptr inbounds ([7 x i8]* @str26, i64 0, i64 0))
tail call void @free(i8* %call) #4
tail call void @free(i8* %call7) #4
tail call void @free(i8* %call12) #4
ret i32 0
}
; Function Attrs: nounwind
declare void @srand(i32) #1
; Function Attrs: nounwind
declare i32 @fprintf(%struct._IO_FILE* nocapture, i8* nocapture, ...) #1
; Function Attrs: noreturn nounwind
declare void @exit(i32) #5
declare float @fabsf(float)
; Function Attrs: nounwind
declare i32 @puts(i8* nocapture) #4
attributes #0 = { nounwind uwtable "less-precise-fpmad"="false" "no-frame-pointer-elim"="false" "no-frame-pointer-elim-non-leaf"="false" "no-infs-fp-math"="false" "no-nans-fp-math"="false" "unsafe-fp-math"="false" "use-soft-float"="false" }
attributes #1 = { nounwind "less-precise-fpmad"="false" "no-frame-pointer-elim"="false" "no-frame-pointer-elim-non-leaf"="false" "no-infs-fp-math"="false" "no-nans-fp-math"="false" "unsafe-fp-math"="false" "use-soft-float"="false" }
attributes #2 = { nounwind readnone uwtable "less-precise-fpmad"="false" "no-frame-pointer-elim"="false" "no-frame-pointer-elim-non-leaf"="false" "no-infs-fp-math"="false" "no-nans-fp-math"="false" "unsafe-fp-math"="false" "use-soft-float"="false" }
attributes #3 = { "less-precise-fpmad"="false" "no-frame-pointer-elim"="false" "no-frame-pointer-elim-non-leaf"="false" "no-infs-fp-math"="false" "no-nans-fp-math"="false" "unsafe-fp-math"="false" "use-soft-float"="false" }
attributes #4 = { nounwind }
attributes #5 = { noreturn nounwind "less-precise-fpmad"="false" "no-frame-pointer-elim"="false" "no-frame-pointer-elim-non-leaf"="false" "no-infs-fp-math"="false" "no-nans-fp-math"="false" "unsafe-fp-math"="false" "use-soft-float"="false" }
attributes #6 = { nounwind readnone "less-precise-fpmad"="false" "no-frame-pointer-elim"="false" "no-frame-pointer-elim-non-leaf"="false" "no-infs-fp-math"="false" "no-nans-fp-math"="false" "unsafe-fp-math"="false" "use-soft-float"="false" }
attributes #7 = { noreturn nounwind }
!0 = metadata !{metadata !"float", metadata !1}
!1 = metadata !{metadata !"omnipotent char", metadata !2}
!2 = metadata !{metadata !"Simple C/C++ TBAA"}
!3 = metadata !{metadata !"long", metadata !1}
!4 = metadata !{metadata !"int", metadata !1}
!5 = metadata !{metadata !"any pointer", metadata !1}
0% or .
You are about to add 0 people to the discussion. Proceed with caution.
Finish editing this message first!
Please register or to comment