From cd8a3dcb4ca42f22ad8f46a95e00977c92be6bbd Mon Sep 17 00:00:00 2001
From: AlexeyAB <alexeyab84@gmail.com>
Date: Thu, 08 Feb 2018 23:22:42 +0000
Subject: [PATCH] Compile fixes
---
src/network_kernels.cu | 13 ++++++++++---
1 files changed, 10 insertions(+), 3 deletions(-)
diff --git a/src/network_kernels.cu b/src/network_kernels.cu
index 9c431cf..64f4f9b 100644
--- a/src/network_kernels.cu
+++ b/src/network_kernels.cu
@@ -51,6 +51,8 @@
fill_ongpu(l.outputs * l.batch, 0, l.delta_gpu, 1);
}
l.forward_gpu(l, state);
+ if(net.wait_stream)
+ cudaStreamSynchronize(get_cuda_stream());
state.input = l.output_gpu;
}
}
@@ -64,6 +66,7 @@
for(i = net.n-1; i >= 0; --i){
state.index = i;
layer l = net.layers[i];
+ if (l.stopbackward) break;
if(i == 0){
state.input = original_input;
state.delta = original_delta;
@@ -78,11 +81,13 @@
void update_network_gpu(network net)
{
+ cuda_set_device(net.gpu_index);
int i;
int update_batch = net.batch*net.subdivisions;
float rate = get_current_rate(net);
for(i = 0; i < net.n; ++i){
layer l = net.layers[i];
+ l.t = get_current_batch(net);
if(l.update_gpu){
l.update_gpu(l, update_batch, rate, net.momentum, net.decay);
}
@@ -109,6 +114,7 @@
state.truth = *net.truth_gpu;
state.train = 1;
forward_network_gpu(net, state);
+ cudaStreamSynchronize(get_cuda_stream());
backward_network_gpu(net, state);
}
@@ -134,7 +140,6 @@
free(ptr);
cuda_set_device(args.net.gpu_index);
*args.err = train_network(args.net, args.d);
- printf("%d\n", args.net.gpu_index);
return 0;
}
@@ -177,6 +182,7 @@
{
int update_batch = net.batch*net.subdivisions;
float rate = get_current_rate(net);
+ l.t = get_current_batch(net);
if(l.update_gpu){
l.update_gpu(l, update_batch, rate, net.momentum, net.decay);
}
@@ -376,7 +382,7 @@
float *get_network_output_layer_gpu(network net, int i)
{
layer l = net.layers[i];
- cuda_pull_array(l.output_gpu, l.output, l.outputs*l.batch);
+ if(l.type != REGION) cuda_pull_array(l.output_gpu, l.output, l.outputs*l.batch);
return l.output;
}
@@ -389,7 +395,8 @@
float *network_predict_gpu(network net, float *input)
{
- cuda_set_device(net.gpu_index);
+ if (net.gpu_index != cuda_get_device())
+ cuda_set_device(net.gpu_index);
int size = get_network_input_size(net) * net.batch;
network_state state;
state.index = 0;
--
Gitblit v1.10.0