debuging

IR Source Level Advanced Debuging

quick-make.sh 脚本

gdb 高级使用

在熟悉了gdb的高级使用之后, 就基本上不再需要下面的 ValueProfiling了.

cgpop 日志

一些无法消除的函数见下:

__mpi2s_gshalo_MOD_mpi2s_gshalo_init

1. call void bitcast (void (...)* @mpi_allreduce_ to void (i8*, i8*, i32*, i32*, i32*, i32*, i32*)*)(i8* %39, i8* %58, i32* @__mpi2s_gshalo_MOD_nprocs, i32* @121, i32* @122, i32* noalias %comm, i32* %ierr) #1 in block:"139"

sCount=0
do i=1,nRecv
   sCount(rNeigh(i)) = 1
enddo
call MPI_Allreduce(sCount,sCount2,nprocs,MPI_INTEGER,MPI_SUM,COMM,ierr)

nSend=sCount2(my_task)

由源代码来看, 是无法消除的, 因为后面的循环次数依赖 sCount2 .

1. call void bitcast (void (...)* @mpi_irecv_ to void (i32*, i32*, i32*, i32*, i32*, i32*, i32*, i32*)*)(i32* %550, i32* %len, i32* @128, i32* %src, i32* %tag, i32* noalias %comm, i32* %547, i32* %ierr) #1 in block:"479"

do i=1,nSend
   len = SendCnt(i)
   iptr = ptrSend(i)
   src  = sNeigh(i)
   call MPI_Irecv(rtmpBuf(iptr),len,MPI_INTEGER,src, &
       tag, COMM,Srequest(i),ierr)
enddo

do i=1,lenSendBuffer
   ig = rtmpBuf(i)
   call LinearOrderedFind(LinearGdof(1:nActive),ig,found,idx)
   halo2send(i) = idx
enddo

由代码看来, 该函数也无法消除. 其参数 rtmpBuf 会赋值给 ig, 而该赋值语句 之后会传递给 LinearOrderedFind 函数.

在编译完成后, 为了检查是否是异常退出, 可以用gdb 捕获 __exit_mod_MOD_exit_pop 函数.

ValueProfiling的基本使用

众所周知, 直接调试IR级别的源代码是非常困难的, 因此我们创造了ValueProfling工具来 帮助完成这项工作, 下面介绍如何深度的使用ValueProfiling来辅助我们手工的定位由于 优化过度而出错的地方

在 cookbook.rest 和 loops.rest 中已经介绍了, 如何使用ValueProfling:

opt -load src/libLLVMPred.so -Insert-Trip-Count -insert-value-profiling

具体来说, 需要一个Pass来告诉ValueProfling捕获哪些 感兴趣 的值, Insert-Trip-Count 就是关心所有Loop的循环次数. 这样就完成了插桩了.

ValueProfiling的进阶使用

有时候我们可能只关心某些值的内容, 例如跟踪一个数据的变化过程, 确定是哪一步开始 出现了偏差, 从而好定位到那个地方, 阅读IR找出问题. 这个时候就需要手工插桩了.

ValueProfiling 的准备有两种方法:

1. 借助前面的命令, 让它给我们 搭建 好ValueProfling的框架环境, ValueProfling 需要 一个 @ValueProfCounters = internal global [? x i32] zeroinitializer 来记录 一共有多少条捕获调用 ? 的地方需要替换成具体的数字.然后每个捕获都是类似于 call void @llvm_profiling_trap_value(i32 ?, i32 %2, i32 0) , ? 的地方是 对应的id. 因此也就是说, 如果我们手工插入call的话, 需要修改这两个地方保持一致.

2. 完全手动插桩, 直接在需要捕获的地方手写 call void @llvm_profiling_trap_value(i32 0, i32 %2, i32 0) , id 从 0 开始增加, 然后随便 某个地方写入 declare void @llvm_profiling_trap_value(i32, i32, i32). 对ir修 改完成之后使用 opt -insert-value-profling 由于没有一个Pass指定捕获哪些变量, 会在整个module中搜寻那些我们手写的 trap_value 函数, 并自动补充其它缺失的部分.

   由于这个时候混合了两种桩, Perf 和 Value, Perf 要求使用插桩前的 bitcode, 这个 时候就看不到Value的结果. Value 要求使用插桩后的 bitcode, 这个时候 Perf 的信息 就是错误的. 需要特别注意.

下面是一个示例: 我们需要确定NAS的cg程序中sparse函数里面跟踪%rowstr这个变量, 因 为它在我们使用 Reduce 优化之后出现了异常, 里面的数据不对. 很明显是变量的初 始化部分出错了, 因此找到变量的初始化主要在sparse函数中.

首先找到前两条关于 %rowstr 的指令是:

%9 = getelementptr [0 x i32]* %rowstr, i64 0, i64 %8
store i32 0, i32* %9, align 4

明显, store的地方不需要捕获, 捕获了也是0, 也许 %8 有捕获的价值, 但这里就跳过了.

接下来找到:

%30 = getelementptr [0 x i32]* %rowstr, i64 0, i64 %29
%31 = load i32* %30, align 4

明显我们可以捕获 %31, 看看 %rowstr里面的内容, 因此修改成下面的样子:

%30 = getelementptr [0 x i32]* %rowstr, i64 0, i64 %29
%31 = load i32* %30, align 4
call void @llvm_profiling_trap_value(i32 54, i32 %31, i32 0)

由于自动插入的是id为0-53的捕获, 因此这里trap_value的id保持继续自增就行了.

最后修改完, id变成59了, 因此再回过头来修改计数器和初始化调用:

@ValueProfCounters = internal global [60 x i32] zeroinitializer
%newargc = call i32 @llvm_start_value_profiling(i32 %argc, i8** %argv, i32* \
   getelementptr inbounds ([60 x i32]* @ValueProfCounters, i64 0, i64 0), i32 \
   60)

对优化后的IR的修改就完成了, 然后再同样的去修改优化前的IR, 两个都拿去编译运行, 把跑出来的结果进行对比, 就能够发现rowstr是从哪里开始出现偏差的了.