 
         d_architecture.init(architecture, num_layers + 1);
 
-        d_weights_storage.resize(num_layers);
-        weights_vec.resize(num_layers);
+        size_t total_weights_size = 0;
+        size_t total_bias_size = 0;
 
         for (unsigned l = 0;  l < num_layers;  ++l) {
             int ni = architecture[l];
+            int no = architecture[l + 1];
             int w_stride = w_strides[l];
-            d_weights_storage[l].init(weights[l], ni * w_stride);
-            weights_vec[l] = d_weights_storage[l];
-
-            //cerr << "layer " << l << ": no = " << no << " w_stride = "
-            //     << w_stride << endl;
-        }
-    
-        d_weights.init(&weights_vec[0], num_layers);
-
-        d_biases_storage.resize(num_layers);
-        biases_vec.resize(num_layers);
-
+            total_weights_size += ni * w_stride;
+            total_bias_size += no;
+            // TODO: align?
+        }
+
+        d_weights.init(total_weights_size);
+        d_biases.init(total_bias_size);
+        
+        // Now copy them all in
+
+        size_t weights_start_offset = 0;
+        size_t bias_start_offset = 0;
+        
         for (unsigned l = 0;  l < num_layers;  ++l) {
+            int ni = architecture[l];
             int no = architecture[l + 1];
-            d_biases_storage[l].init(biases[l], no);
-            biases_vec[l] = d_biases_storage[l];
-        }
-    
-        d_biases.init(&biases_vec[0], num_layers);
+            int w_stride = w_strides[l];
+            size_t w_size = ni * w_stride;
+
+            cudaError_t err
+                = cudaMemcpy(d_weights + weights_start_offset,
+                             weights[l],
+                             w_size * sizeof(float),
+                             cudaMemcpyHostToDevice);
+            
+            if (err != cudaSuccess)
+                throw Exception(cudaGetErrorString(err));
+
+            err = cudaMemcpy(d_biases + bias_start_offset,
+                             biases[l],
+                             no * sizeof(float),
+                             cudaMemcpyHostToDevice);
+            
+            if (err != cudaSuccess)
+                throw Exception(cudaGetErrorString(err));
+            
+            weights_start_offset += ni * w_stride;
+            bias_start_offset += no;
+            // TODO: align?
+        }
 
         d_w_strides.init(w_strides, num_layers);
         


            __syncthreads();

            if (tid < no) scratch[tid] = biases[l][tid];

            __syncthreads();
            
            /* Now we accumulate them, allowing each thread to increment in its

    }
}


/** Train a fully-connected neural network architecture via backpropagation
    one a single training example.  The work is split over all of the cores
