From 1c05ebf522f0bb5776ba51a46d94aa101220fea1 Mon Sep 17 00:00:00 2001
From: AlexeyAB <alexeyab84@gmail.com>
Date: Thu, 07 Jun 2018 00:39:30 +0000
Subject: [PATCH] Minor fix
---
src/convolutional_layer.c | 20 +++++++++++++++++---
1 files changed, 17 insertions(+), 3 deletions(-)
diff --git a/src/convolutional_layer.c b/src/convolutional_layer.c
index cd36929..90f9551 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;
}
@@ -460,6 +465,10 @@
l->x_norm = realloc(l->x_norm, l->batch*l->outputs*sizeof(float));
}
+ if (l->xnor) {
+ //l->binary_input = realloc(l->inputs*l->batch, sizeof(float));
+ }
+
#ifdef GPU
if (old_w < w || old_h < h) {
cuda_free(l->delta_gpu);
@@ -475,6 +484,11 @@
l->x_gpu = cuda_make_array(l->output, l->batch*l->outputs);
l->x_norm_gpu = cuda_make_array(l->output, l->batch*l->outputs);
}
+
+ if (l->xnor) {
+ cuda_free(l->binary_input_gpu);
+ l->binary_input_gpu = cuda_make_array(0, l->inputs*l->batch);
+ }
}
#ifdef CUDNN
cudnn_convolutional_setup(l, cudnn_fastest);
@@ -488,7 +502,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: %zu, available: %zu\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