| | |
| | | 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 dimGrid((size-1)/BLOCK + 1, n, batch); |
| | |
| | | } |
| | | } |
| | | |
| | | 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); |
| | | 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 m = layer.n; |
| | |
| | | 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 m = layer.n; |
| | |
| | | } |
| | | } |
| | | |
| | | 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.biases_gpu, layer.biases, 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.biases_gpu, layer.biases, 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; |
| | | |
| | |
| | | #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 |
| | | void draw_detection(image im, float *box, int side) |
| | | { |
| | | int classes = 20; |
| | | int classes = 21; |
| | | int elems = 4+classes; |
| | | int j; |
| | | int r, c; |
| | |
| | | if(weightfile){ |
| | | load_weights(&net, weightfile); |
| | | } |
| | | net.seen = 0; |
| | | printf("Learning Rate: %g, Momentum: %g, Decay: %g\n", net.learning_rate, net.momentum, net.decay); |
| | | int imgs = 128; |
| | | srand(time(0)); |
| | |
| | | int im_dim = 512; |
| | | int jitter = 64; |
| | | 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); |
| | | clock_t time; |
| | | while(1){ |
| | |
| | | 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]); |
| | | draw_detection(im, train.y.vals[0], 7); |
| | | image im = float_to_image(im_dim - jitter, im_dim-jitter, 3, train.X.vals[114]); |
| | | draw_detection(im, train.y.vals[114], 7); |
| | | show_image(im, "truth"); |
| | | cvWaitKey(0); |
| | | */ |
| | |
| | | char **paths = (char **)list_to_array(plist); |
| | | int im_size = 448; |
| | | int classes = 20; |
| | | int background = 1; |
| | | int background = 0; |
| | | int num_output = 7*7*(4+classes+background); |
| | | |
| | | int m = plist->size; |
| | |
| | | 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; |
| | | int size = layer.inputs*layer.batch; |
| | |
| | | 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; |
| | | int size = layer.inputs*layer.batch; |
| | |
| | | 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) |
| | | { |
| | | float *out = get_network_output(net); |
| | |
| | | #include "dropout_layer.h" |
| | | } |
| | | |
| | | extern "C" 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_output_gpu_layer(network net, int i); |
| | | float * get_network_delta_gpu_layer(network net, int i); |
| | | float *get_network_output_gpu(network net); |
| | | |
| | | void forward_network_gpu(network net, network_state state) |
| | |
| | | state.train = 1; |
| | | forward_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); |
| | | if ((net.seen / net.batch) % net.subdivisions == 0) update_network_gpu(net); |
| | | |
| | | return error; |
| | | } |