From ced198e9390195875d743d77eadece99c7fd5b38 Mon Sep 17 00:00:00 2001
From: AlexeyAB <alexeyab84@gmail.com>
Date: Mon, 19 Mar 2018 23:17:26 +0000
Subject: [PATCH] Fixed gpu_id for DLL/SO
---
src/network.c | 84 ++++++++++++++++++++++++++++++++++--------
1 files changed, 68 insertions(+), 16 deletions(-)
diff --git a/src/network.c b/src/network.c
index 01b7962..61f87c5 100644
--- a/src/network.c
+++ b/src/network.c
@@ -41,7 +41,7 @@
net.momentum = 0;
net.decay = 0;
#ifdef GPU
- if(gpu_index >= 0) update_network_gpu(net);
+ //if(net.gpu_index >= 0) update_network_gpu(net);
#endif
}
@@ -50,6 +50,7 @@
int batch_num = get_current_batch(net);
int i;
float rate;
+ if (batch_num < net.burn_in) return net.learning_rate * pow((float)batch_num / net.burn_in, net.power);
switch (net.policy) {
case CONSTANT:
return net.learning_rate;
@@ -60,14 +61,15 @@
for(i = 0; i < net.num_steps; ++i){
if(net.steps[i] > batch_num) return rate;
rate *= net.scales[i];
- if(net.steps[i] > batch_num - 1) reset_momentum(net);
+ //if(net.steps[i] > batch_num - 1 && net.scales[i] > 1) reset_momentum(net);
}
return rate;
case EXP:
return net.learning_rate * pow(net.gamma, batch_num);
case POLY:
- if (batch_num < net.burn_in) return net.learning_rate * pow((float)batch_num / net.burn_in, net.power);
- return net.learning_rate * pow(1 - (float)batch_num / net.max_batches, net.power);
+ return net.learning_rate * pow(1 - (float)batch_num / net.max_batches, net.power);
+ //if (batch_num < net.burn_in) return net.learning_rate * pow((float)batch_num / net.burn_in, net.power);
+ //return net.learning_rate * pow(1 - (float)batch_num / net.max_batches, net.power);
case RANDOM:
return net.learning_rate * pow(rand_uniform(0,1), net.power);
case SIG:
@@ -138,6 +140,11 @@
#ifdef GPU
net.input_gpu = calloc(1, sizeof(float *));
net.truth_gpu = calloc(1, sizeof(float *));
+
+ net.input16_gpu = calloc(1, sizeof(float *));
+ net.output16_gpu = calloc(1, sizeof(float *));
+ net.max_input16_size = calloc(1, sizeof(size_t));
+ net.max_output16_size = calloc(1, sizeof(size_t));
#endif
return net;
}
@@ -218,6 +225,7 @@
state.delta = prev.delta;
}
layer l = net.layers[i];
+ if (l.stopbackward) break;
l.backward(l, state);
}
}
@@ -313,7 +321,20 @@
net->layers[i].batch = b;
#ifdef CUDNN
if(net->layers[i].type == CONVOLUTIONAL){
- cudnn_convolutional_setup(net->layers + i);
+ cudnn_convolutional_setup(net->layers + i, cudnn_fastest);
+ /*
+ layer *l = net->layers + i;
+ cudnn_convolutional_setup(l, cudnn_fastest);
+ // check for excessive memory consumption
+ size_t free_byte;
+ 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");
+ cudnn_convolutional_setup(l, cudnn_smallest);
+ l->workspace_size = get_workspace_size(*l);
+ }
+ */
}
#endif
}
@@ -321,6 +342,18 @@
int resize_network(network *net, int w, int h)
{
+#ifdef GPU
+ cuda_set_device(net->gpu_index);
+ if(gpu_index >= 0){
+ cuda_free(net->workspace);
+ if (net->input_gpu) {
+ cuda_free(*net->input_gpu);
+ *net->input_gpu = 0;
+ cuda_free(*net->truth_gpu);
+ *net->truth_gpu = 0;
+ }
+ }
+#endif
int i;
//if(w == net->w && h == net->h) return 0;
net->w = w;
@@ -331,12 +364,19 @@
//fflush(stderr);
for (i = 0; i < net->n; ++i){
layer l = net->layers[i];
+ //printf(" %d: layer = %d,", i, l.type);
if(l.type == CONVOLUTIONAL){
resize_convolutional_layer(&l, w, h);
}else if(l.type == CROP){
resize_crop_layer(&l, w, h);
}else if(l.type == MAXPOOL){
resize_maxpool_layer(&l, w, h);
+ }else if(l.type == REGION){
+ resize_region_layer(&l, w, h);
+ }else if(l.type == ROUTE){
+ resize_route_layer(&l, net);
+ }else if (l.type == SHORTCUT) {
+ resize_shortcut_layer(&l, w, h);
}else if(l.type == REORG){
resize_reorg_layer(&l, w, h);
}else if(l.type == AVGPOOL){
@@ -346,6 +386,7 @@
}else if(l.type == COST){
resize_cost_layer(&l, inputs);
}else{
+ fprintf(stderr, "Resizing type %d \n", (int)l.type);
error("Cannot resize this type of layer");
}
if(l.workspace_size > workspace_size) workspace_size = l.workspace_size;
@@ -357,8 +398,9 @@
}
#ifdef GPU
if(gpu_index >= 0){
- cuda_free(net->workspace);
+ printf(" try to allocate workspace = %zu * sizeof(float), ", (workspace_size - 1) / sizeof(float) + 1);
net->workspace = cuda_make_array(0, (workspace_size-1)/sizeof(float)+1);
+ printf(" CUDA allocate done! \n");
}else {
free(net->workspace);
net->workspace = calloc(1, workspace_size);
@@ -565,7 +607,6 @@
return acc;
}
-
float network_accuracy_multi(network net, data d, int n)
{
matrix guess = network_predict_data_multi(net, d, n);
@@ -576,15 +617,26 @@
void free_network(network net)
{
- int i;
- for(i = 0; i < net.n; ++i){
- free_layer(net.layers[i]);
- }
- free(net.layers);
+ int i;
+ for (i = 0; i < net.n; ++i) {
+ free_layer(net.layers[i]);
+ }
+ free(net.layers);
#ifdef GPU
- if(*net.input_gpu) cuda_free(*net.input_gpu);
- if(*net.truth_gpu) cuda_free(*net.truth_gpu);
- if(net.input_gpu) free(net.input_gpu);
- if(net.truth_gpu) free(net.truth_gpu);
+ if (gpu_index >= 0) cuda_free(net.workspace);
+ else free(net.workspace);
+ if (*net.input_gpu) cuda_free(*net.input_gpu);
+ if (*net.truth_gpu) cuda_free(*net.truth_gpu);
+ if (net.input_gpu) free(net.input_gpu);
+ if (net.truth_gpu) free(net.truth_gpu);
+
+ if (*net.input16_gpu) cuda_free(*net.input16_gpu);
+ if (*net.output16_gpu) cuda_free(*net.output16_gpu);
+ if (net.input16_gpu) free(net.input16_gpu);
+ if (net.output16_gpu) free(net.output16_gpu);
+ if (net.max_input16_size) free(net.max_input16_size);
+ if (net.max_output16_size) free(net.max_output16_size);
+#else
+ free(net.workspace);
#endif
}
--
Gitblit v1.10.0