OrgLion-ML
/
darknet
mirror of https://github.com/AlexeyAB/darknet.git
1
0
Fork 0
Code Issues Projects Releases Wiki Activity

going to break stuff

This commit is contained in:
Joseph Redmon 2015-03-22 21:28:45 -07:00
parent 664c5dd2f2
commit 7100de0b59
5 changed files with 21 additions and 38 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

14
src/convolutional_kernels.cu
View File

@ -17,7 +17,7 @@ __global__ void bias_output_kernel(float *output, float *biases, int n, int size
if(offset < size) output[(batch*n+filter)*size + offset] = biases[filter]; if(offset < size) output[(batch*n+filter)*size + offset] = biases[filter];
} }
extern "C" void bias_output_gpu(float *output, float *biases, int batch, int n, int size) void bias_output_gpu(float *output, float *biases, int batch, int n, int size)
{ {
dim3 dimBlock(BLOCK, 1, 1); dim3 dimBlock(BLOCK, 1, 1);
dim3 dimGrid((size-1)/BLOCK + 1, n, batch); dim3 dimGrid((size-1)/BLOCK + 1, n, batch);
@ -46,13 +46,13 @@ __global__ void backward_bias_kernel(float *bias_updates, float *delta, int batc
} }
} }
extern "C" void backward_bias_gpu(float *bias_updates, float *delta, int batch, int n, int size) void backward_bias_gpu(float *bias_updates, float *delta, int batch, int n, int size)
{ {
backward_bias_kernel<<<n, BLOCK>>>(bias_updates, delta, batch, n, size, 1); backward_bias_kernel<<<n, BLOCK>>>(bias_updates, delta, batch, n, size, 1);
check_error(cudaPeekAtLastError()); check_error(cudaPeekAtLastError());
} }
extern "C" void forward_convolutional_layer_gpu(convolutional_layer layer, network_state state) void forward_convolutional_layer_gpu(convolutional_layer layer, network_state state)
{ {
int i; int i;
int m = layer.n; int m = layer.n;
@ -71,7 +71,7 @@ extern "C" void forward_convolutional_layer_gpu(convolutional_layer layer, netwo
activate_array_ongpu(layer.output_gpu, m*n*layer.batch, layer.activation); activate_array_ongpu(layer.output_gpu, m*n*layer.batch, layer.activation);
} }
extern "C" void backward_convolutional_layer_gpu(convolutional_layer layer, network_state state) void backward_convolutional_layer_gpu(convolutional_layer layer, network_state state)
{ {
int i; int i;
int m = layer.n; int m = layer.n;
@ -105,7 +105,7 @@ extern "C" void backward_convolutional_layer_gpu(convolutional_layer layer, netw
} }
} }
extern "C" void pull_convolutional_layer(convolutional_layer layer) void pull_convolutional_layer(convolutional_layer layer)
{ {
cuda_pull_array(layer.filters_gpu, layer.filters, layer.c*layer.n*layer.size*layer.size); cuda_pull_array(layer.filters_gpu, layer.filters, layer.c*layer.n*layer.size*layer.size);
cuda_pull_array(layer.biases_gpu, layer.biases, layer.n); cuda_pull_array(layer.biases_gpu, layer.biases, layer.n);
@ -113,7 +113,7 @@ extern "C" void pull_convolutional_layer(convolutional_layer layer)
cuda_pull_array(layer.bias_updates_gpu, layer.bias_updates, layer.n); cuda_pull_array(layer.bias_updates_gpu, layer.bias_updates, layer.n);
} }
extern "C" void push_convolutional_layer(convolutional_layer layer) void push_convolutional_layer(convolutional_layer layer)
{ {
cuda_push_array(layer.filters_gpu, layer.filters, layer.c*layer.n*layer.size*layer.size); cuda_push_array(layer.filters_gpu, layer.filters, layer.c*layer.n*layer.size*layer.size);
cuda_push_array(layer.biases_gpu, layer.biases, layer.n); cuda_push_array(layer.biases_gpu, layer.biases, layer.n);
@ -121,7 +121,7 @@ extern "C" void push_convolutional_layer(convolutional_layer layer)
cuda_push_array(layer.bias_updates_gpu, layer.bias_updates, layer.n); cuda_push_array(layer.bias_updates_gpu, layer.bias_updates, layer.n);
} }
extern "C" void update_convolutional_layer_gpu(convolutional_layer layer, int batch, float learning_rate, float momentum, float decay) void update_convolutional_layer_gpu(convolutional_layer layer, int batch, float learning_rate, float momentum, float decay)
{ {
int size = layer.size*layer.size*layer.c*layer.n; int size = layer.size*layer.size*layer.c*layer.n;

13
src/detection.c
View File

@ -3,11 +3,11 @@
#include "parser.h" #include "parser.h"
char *class_names[] = {"aeroplane", "bicycle", "bird", "boat", "bottle", "bus", "car", "cat", "chair", "cow", "diningtable", "dog", "horse", "motorbike", "person", "pottedplant", "sheep", "sofa", "train", "tvmonitor"}; char *class_names[] = {"bg", "aeroplane", "bicycle", "bird", "boat", "bottle", "bus", "car", "cat", "chair", "cow", "diningtable", "dog", "horse", "motorbike", "person", "pottedplant", "sheep", "sofa", "train", "tvmonitor"};
#define AMNT 3 #define AMNT 3
void draw_detection(image im, float *box, int side) void draw_detection(image im, float *box, int side)
{ {
int classes = 20; int classes = 21;
int elems = 4+classes; int elems = 4+classes;
int j; int j;
int r, c; int r, c;
@ -50,6 +50,7 @@ void train_detection(char *cfgfile, char *weightfile)
if(weightfile){ if(weightfile){
load_weights(&net, weightfile); load_weights(&net, weightfile);
} }
net.seen = 0;
printf("Learning Rate: %g, Momentum: %g, Decay: %g\n", net.learning_rate, net.momentum, net.decay); printf("Learning Rate: %g, Momentum: %g, Decay: %g\n", net.learning_rate, net.momentum, net.decay);
int imgs = 128; int imgs = 128;
srand(time(0)); srand(time(0));
@ -62,7 +63,7 @@ void train_detection(char *cfgfile, char *weightfile)
int im_dim = 512; int im_dim = 512;
int jitter = 64; int jitter = 64;
int classes = 20; int classes = 20;
int background = 1; int background = 0;
pthread_t load_thread = load_data_detection_thread(imgs, paths, plist->size, classes, im_dim, im_dim, 7, 7, jitter, background, &buffer); pthread_t load_thread = load_data_detection_thread(imgs, paths, plist->size, classes, im_dim, im_dim, 7, 7, jitter, background, &buffer);
clock_t time; clock_t time;
while(1){ while(1){
@ -73,8 +74,8 @@ void train_detection(char *cfgfile, char *weightfile)
load_thread = load_data_detection_thread(imgs, paths, plist->size, classes, im_dim, im_dim, 7, 7, jitter, background, &buffer); load_thread = load_data_detection_thread(imgs, paths, plist->size, classes, im_dim, im_dim, 7, 7, jitter, background, &buffer);
/* /*
image im = float_to_image(im_dim - jitter, im_dim-jitter, 3, train.X.vals[0]); image im = float_to_image(im_dim - jitter, im_dim-jitter, 3, train.X.vals[114]);
draw_detection(im, train.y.vals[0], 7); draw_detection(im, train.y.vals[114], 7);
show_image(im, "truth"); show_image(im, "truth");
cvWaitKey(0); cvWaitKey(0);
*/ */
@ -108,7 +109,7 @@ void validate_detection(char *cfgfile, char *weightfile)
char **paths = (char **)list_to_array(plist); char **paths = (char **)list_to_array(plist);
int im_size = 448; int im_size = 448;
int classes = 20; int classes = 20;
int background = 1; int background = 0;
int num_output = 7*7*(4+classes+background); int num_output = 7*7*(4+classes+background);
int m = plist->size; int m = plist->size;

4
src/dropout_layer_kernels.cu
View File

@ -11,7 +11,7 @@ __global__ void yoloswag420blazeit360noscope(float *input, int size, float *rand
if(id < size) input[id] = (rand[id] < prob) ? 0 : input[id]*scale; if(id < size) input[id] = (rand[id] < prob) ? 0 : input[id]*scale;
} }
extern "C" void forward_dropout_layer_gpu(dropout_layer layer, network_state state) void forward_dropout_layer_gpu(dropout_layer layer, network_state state)
{ {
if (!state.train) return; if (!state.train) return;
int size = layer.inputs*layer.batch; int size = layer.inputs*layer.batch;
@ -21,7 +21,7 @@ extern "C" void forward_dropout_layer_gpu(dropout_layer layer, network_state sta
check_error(cudaPeekAtLastError()); check_error(cudaPeekAtLastError());
} }
extern "C" void backward_dropout_layer_gpu(dropout_layer layer, network_state state) void backward_dropout_layer_gpu(dropout_layer layer, network_state state)
{ {
if(!state.delta) return; if(!state.delta) return;
int size = layer.inputs*layer.batch; int size = layer.inputs*layer.batch;

18
src/network.c
View File

@ -194,24 +194,6 @@ float *get_network_delta(network net)
return get_network_delta_layer(net, net.n-1); return get_network_delta_layer(net, net.n-1);
} }
float calculate_error_network(network net, float *truth)
{
float sum = 0;
float *delta = get_network_delta(net);
float *out = get_network_output(net);
int i;
for(i = 0; i < get_network_output_size(net)*net.batch; ++i){
//if(i %get_network_output_size(net) == 0) printf("\n");
//printf("%5.2f %5.2f, ", out[i], truth[i]);
//if(i == get_network_output_size(net)) printf("\n");
delta[i] = truth[i] - out[i];
//printf("%.10f, ", out[i]);
sum += delta[i]*delta[i];
}
//printf("\n");
return sum;
}
int get_predicted_class_network(network net) int get_predicted_class_network(network net)
{ {
float *out = get_network_output(net); float *out = get_network_output(net);

6
src/network_kernels.cu
View File

@ -20,8 +20,8 @@ extern "C" {
#include "dropout_layer.h" #include "dropout_layer.h"
} }
extern "C" float * get_network_output_gpu_layer(network net, int i); float * get_network_output_gpu_layer(network net, int i);
extern "C" float * get_network_delta_gpu_layer(network net, int i); float * get_network_delta_gpu_layer(network net, int i);
float *get_network_output_gpu(network net); float *get_network_output_gpu(network net);
void forward_network_gpu(network net, network_state state) void forward_network_gpu(network net, network_state state)
@ -196,8 +196,8 @@ float train_network_datum_gpu(network net, float *x, float *y)
state.train = 1; state.train = 1;
forward_network_gpu(net, state); forward_network_gpu(net, state);
backward_network_gpu(net, state); backward_network_gpu(net, state);
if ((net.seen / net.batch) % net.subdivisions == 0) update_network_gpu(net);
float error = get_network_cost(net); float error = get_network_cost(net);
if ((net.seen / net.batch) % net.subdivisions == 0) update_network_gpu(net);
return error; return error;
} }