From 1e8781aecdfa7f044d900d6526a1f47b6252b060 Mon Sep 17 00:00:00 2001 From: joydddd Date: Tue, 10 Oct 2023 21:56:18 -0400 Subject: [PATCH] Fix debug functions --- .gitignore | 2 ++ Makefile | 3 ++- gpu/debug.c | 10 +++++++++- gpu/debug.h | 2 +- gpu/gpu.mk | 11 ++++++----- gpu/plchain.cu | 16 +++++++--------- gpu_config.json | 4 ++-- 7 files changed, 29 insertions(+), 19 deletions(-) diff --git a/.gitignore b/.gitignore index c9a9c18d..fb58534d 100644 --- a/.gitignore +++ b/.gitignore @@ -18,3 +18,5 @@ ncu nsys profile_output* workloads +.cmake/** +.depend \ No newline at end of file diff --git a/Makefile b/Makefile index e3d5d03a..6e7ebee1 100644 --- a/Makefile +++ b/Makefile @@ -1,5 +1,5 @@ CFLAGS= -O2 -g -DNDEBUG -CDEBUG_FLAGS= -g -Wall -Wc++-compat -Wextra -O2 #-O0 -DNDEBUG +CDEBUG_FLAGS= -g -Wall -Wc++-compat -Wextra -DDEBUG_PRINT -O2 #-O0 -DNDEBUG CPPFLAGS= -DHAVE_KALLOC -D__AMD_SPLIT_KERNELS__ # -Wno-unused-but-set-variable -Wno-unused-variable INCLUDES= -I . OBJS= kthread.o kalloc.o misc.o bseq.o sketch.o sdust.o options.o index.o \ @@ -40,6 +40,7 @@ ifneq ($(DEBUG),) # turn on debug flags endif ifneq ($(DEBUG_ANALYSIS),) # turn on debug flags CFLAGS = $(CDEBUG_FLAGS) + CFLAGS += -DDEBUG_CHECK -DDEBUG_VERBOSE endif .PHONY:all extra clean depend # profile diff --git a/gpu/debug.c b/gpu/debug.c index 0eebde08..d732ac52 100644 --- a/gpu/debug.c +++ b/gpu/debug.c @@ -33,7 +33,7 @@ void debug_output_anchors(const char debug_folder[], chain_read_t *in) { /* Write Sequence Name and Length, rep_len*/ fprintf(f_anchors, "<%s\t%d\n", in->seq.name, in->seq.len); - fprintf(f_anchors, "*%d\n", in->seq.rep_len); + fprintf(f_anchors, "*%d\n", in->rep_len); /* Read Number of Anchors */ fprintf(f_anchors, "#%d\n", in->n); @@ -45,6 +45,8 @@ void debug_output_anchors(const char debug_folder[], chain_read_t *in) { fprintf(f_anchors, "\n"); } +// DEBUG: not used +#if 0 void debug_output_score(const char debug_folder[], chain_read_t *in) { static FILE *f_score = NULL; if (!f_score) { @@ -73,6 +75,7 @@ void debug_output_score(const char debug_folder[], chain_read_t *in) { } fprintf(f_score, "\n"); } +#endif void debug_output_meta(const char debug_folder[], input_meta_t *meta) { static FILE *f_metaout = NULL; @@ -234,6 +237,10 @@ void debug_print_regs(mm_reg1_t* regs, int n_u, char* qname){ ///////////// check functions /////////////////////////////// /////////////////////////////////////////////////////////////////////////// #ifdef DEBUG_CHECK + + +// DEBUG: uses with gold standard input score and range. SCORE CHECK +#if 0 /** * Read Plaintxt input file for Chaining scores from .score * Allocate and Populate chain_read_t.f, chain_read_t.p @@ -373,6 +380,7 @@ int debug_check_score(const int64_t *p, const int32_t *f, const int64_t *p_gold, readid++; return rt; } +#endif // uses if we have gold standard input void debug_check_range(const int32_t* range, size_t n){ diff --git a/gpu/debug.h b/gpu/debug.h index f04a695b..e92bcb4e 100644 --- a/gpu/debug.h +++ b/gpu/debug.h @@ -7,7 +7,7 @@ extern "C" { #endif -extern char debug_folder[]; +const char debug_folder[] = "debug"; // #define ITER_LIMIT 10000 // #define MAX_READ_NUM 100000 diff --git a/gpu/gpu.mk b/gpu/gpu.mk index dea32fdd..a8f42ec4 100644 --- a/gpu/gpu.mk +++ b/gpu/gpu.mk @@ -11,6 +11,8 @@ CONFIG += $(if $(LONG_CUT),-DMM_LONG_SEG_CUTOFF=\($(LONG_CUT)\)) ################################################### CU_SRC = $(wildcard gpu/*.cu) CU_OBJS = $(CU_SRC:%.cu=%.o) +C_SRC = $(wildcard gpu/*.c) +OBJS += $(C_SRC:%.c=%.o) INCLUDES += -I gpu ################################################### @@ -18,14 +20,14 @@ INCLUDES += -I gpu ################################################### NVCC = nvcc CUDAFLAGS = -rdc=true -lineinfo -CUDATESTFLAG = -G -DDEBUG_PRINT +CUDATESTFLAG = -G ################################################### ############ HIP Compile ################### ################################################### HIPCC = hipcc -HIPFLAGS = -DUSEHIP -Rpass-analysis=kernel-resource-usage -HIPTESTFLAGS = -g -DDEBUG_PRINT +HIPFLAGS = -DUSEHIP +HIPTESTFLAGS = -g -Rpass-analysis=kernel-resource-usage ################################################### ############ DEBUG Options ################### @@ -46,7 +48,6 @@ endif ifneq ($(DEBUG_ANALYSIS),) GPU_FLAGS += $(GPU_TESTFL) - GPU_FLAGS += -DDEBUG_CHECK -DDEBUG_VERBOSE endif @@ -54,7 +55,7 @@ endif $(GPU_CC) -c $(GPU_FLAGS) $(CFLAGS) $(CPPFLAGS) $(INCLUDES) $(CONFIG) $< -o $@ cleangpu: - rm -f $(CU_OBJS) + rm -f gpu/*.o # profile:CFLAGS += -pg -g3 # profile:all diff --git a/gpu/plchain.cu b/gpu/plchain.cu index 8e52a248..6d88f644 100644 --- a/gpu/plchain.cu +++ b/gpu/plchain.cu @@ -426,11 +426,13 @@ void plchain_debug_analysis(stream_ptr_t stream){ fprintf(stderr, "[DEBUG] total segs: %lu, short:%lu mid: %u long: %u\n", cut_num, cut_num - num_mid_seg - num_long_seg, num_mid_seg, num_long_seg); -// DEBUG: check range w.r.t to input and range violations -#if defined(DEBUG_CHECK) && 0 + int32_t* range = (int32_t*)malloc(sizeof(int32_t) * total_n); cudaMemcpy(range, dev_mem->d_range, sizeof(int32_t) * total_n, cudaMemcpyDeviceToHost); + size_t* cut = (size_t*)malloc(sizeof(size_t) * cut_num); + cudaMemcpy(cut, dev_mem->d_cut, sizeof(size_t) * cut_num, + cudaMemcpyDeviceToHost); // Check range w.r.t input (MAKE SURE INPUT RANGE EXISTS) #if 0 @@ -447,22 +449,18 @@ void plchain_debug_analysis(stream_ptr_t stream){ // DEBUG: Check voilation of cut #if defined(DEBUG_CHECK) && 0 - size_t* cut = (size_t*)malloc(sizeof(size_t) * cut_num); - cudaMemcpy(cut, dev_mem->d_cut, sizeof(size_t) * cut_num, - cudaMemcpyDeviceToHost); for (int readid = 0, cid = 0, idx = 0; readid < dev_mem->size; readid++) { // DEBUG: Print cuts #if defined(DEBUG_VERBOSE) && 0 debug_print_cut(cut + cid, cut_num - cid, reads[readid].n, idx, reads[readid].seq.name); #endif - cid += debug_check_cut(cut + cid, range, cut_num - cid, - reads[readid].n, idx); + cid += debug_check_cut(cut + cid, range, cut_num - cid, reads[readid].n, idx); idx += reads[readid].n; } - free(cut); #endif + + free(cut); free(range); -#endif // DEBUG_CHECK // DEBUG: Calculate workload distribution #if defined(DEBUG_VERBOSE) && 1 diff --git a/gpu_config.json b/gpu_config.json index 50869b30..1f0778a0 100644 --- a/gpu_config.json +++ b/gpu_config.json @@ -2,8 +2,8 @@ "num_streams": 1, "min_n": 512, "//min_n": "queries with less anchors will be handled on cpu", - "max_total_n": 126844000, - "max_read": 8120009, + "//max_total_n": 126844000, + "//max_read": 8120009, "avg_read_n": 20000, "//avg_read_n": "expect average number of anchors per read", "range_kernel": {