From 8b9af80a5b3bdb6ec17c4fa2cfa14ce4676ef1c3 Mon Sep 17 00:00:00 2001
From: Alexey <AlexeyAB@users.noreply.github.com>
Date: Fri, 18 May 2018 20:27:59 +0000
Subject: [PATCH] Merge pull request #851 from NadeenUdantha/patch-1
---
src/convolutional_layer.c | 11 ++++++++---
1 files changed, 8 insertions(+), 3 deletions(-)
diff --git a/src/convolutional_layer.c b/src/convolutional_layer.c
index cd36929..a349454 100644
--- a/src/convolutional_layer.c
+++ b/src/convolutional_layer.c
@@ -178,6 +178,8 @@
// batch norm
cudnnSetTensor4dDescriptor(l->normTensorDesc, CUDNN_TENSOR_NCHW, CUDNN_DATA_FLOAT, 1, l->out_c, 1, 1);
cudnnSetTensor4dDescriptor(l->normDstTensorDesc, CUDNN_TENSOR_NCHW, CUDNN_DATA_FLOAT, l->batch, l->out_c, l->out_h, l->out_w);
+
+ cudnnSetTensor4dDescriptor(l->normDstTensorDescF16, CUDNN_TENSOR_NCHW, data_type, l->batch, l->out_c, l->out_h, l->out_w);
#if(CUDNN_MAJOR >= 6)
cudnnSetConvolution2dDescriptor(l->convDesc, l->pad, l->pad, l->stride, l->stride, 1, 1, CUDNN_CROSS_CORRELATION, CUDNN_DATA_FLOAT); // cudnn >= 6.0
#else
@@ -248,7 +250,7 @@
//printf("Tensor Cores - Backward-filter enabled: l->bf_algo = CUDNN_CONVOLUTION_BWD_FILTER_ALGO_WINOGRAD_NONFUSED \n");
if (fw == 2 && bd == 2 && bf == 2) printf("TF ");
- else if (fw >= 1 && bd >= 1 && bf >= 1) printf("TH ");
+ else if (fw == 1 && bd == 1 && bf == 1) printf("TH ");
}
}
#endif
@@ -379,6 +381,7 @@
}
#ifdef CUDNN
cudnnCreateTensorDescriptor(&l.normDstTensorDesc);
+ cudnnCreateTensorDescriptor(&l.normDstTensorDescF16);
cudnnCreateTensorDescriptor(&l.normTensorDesc);
cudnnCreateTensorDescriptor(&l.srcTensorDesc);
cudnnCreateTensorDescriptor(&l.dstTensorDesc);
@@ -394,7 +397,9 @@
l.workspace_size = get_workspace_size(l);
l.activation = activation;
- fprintf(stderr, "conv %5d %2d x%2d /%2d %4d x%4d x%4d -> %4d x%4d x%4d\n", n, size, size, stride, w, h, c, l.out_w, l.out_h, l.out_c);
+ //fprintf(stderr, "conv %5d %2d x%2d /%2d %4d x%4d x%4d -> %4d x%4d x%4d\n", n, size, size, stride, w, h, c, l.out_w, l.out_h, l.out_c);
+ l.bflops = (2.0 * l.n * l.size*l.size*l.c * l.out_h*l.out_w) / 1000000000.;
+ fprintf(stderr, "conv %5d %2d x%2d /%2d %4d x%4d x%4d -> %4d x%4d x%4d %5.3f BF\n", n, size, size, stride, w, h, c, l.out_w, l.out_h, l.out_c, l.bflops);
return l;
}
@@ -488,7 +493,7 @@
size_t total_byte;
check_error(cudaMemGetInfo(&free_byte, &total_byte));
if (l->workspace_size > free_byte || l->workspace_size >= total_byte / 2) {
- printf(" used slow CUDNN algo without Workspace! \n");
+ printf(" used slow CUDNN algo without Workspace! Need memory: %d, available: %d\n", l->workspace_size, (free_byte < total_byte/2) ? free_byte : total_byte/2);
cudnn_convolutional_setup(l, cudnn_smallest);
l->workspace_size = get_workspace_size(*l);
}
--
Gitblit v1.10.0