   return mean;
 }
 
+template<typename T>
+void GPU_remove_baseline(T* d_collection, int nsamps){
+  float mean  = 0.0;
+
+  int count = 0;
+  do{
+    mean = GPU_mean(d_collection, nsamps, 0);
+
+    thrust::for_each(thrust::device_ptr<T>(d_collection), 
+        thrust::device_ptr<T>(d_collection)+nsamps,thrust::placeholders::_1 -= mean);
+
+    std::cout << "mean: " << mean <<std::endl;
+
+    if(count++ > 100) break;
+
+  }while(abs(mean) > 5e-7 * nsamps);
+
+  thrust::for_each(thrust::device_ptr<T>(d_collection), 
+        thrust::device_ptr<T>(d_collection)+nsamps,thrust::placeholders::_1 -= mean);
+
+
+//  float mean = GPU_mean(d_collection, nsamps, 0);
+
+//  thrust::for_each(thrust::device_ptr<T>(d_collection),   thrust::device_ptr<T>(d_collection)+nsamps,thrust::placeholders::_1 -= mean);
+
+}
+
 
 template<typename T>
 void GPU_fill(T* start, T* end, T value){

template void GPU_fill<float>(float*, float*, float);
template float GPU_rms<float>(float*,int,int);
template float GPU_mean<float>(float*,int,int);

__global__
void normalisation_kernel(float*d_powers, float mean, float sigma, 

template void device_conversion<char,float>(char*, float*, unsigned int, unsigned int, unsigned int);
template void device_conversion<unsigned char,float>(unsigned char*, float*, unsigned int, unsigned int, unsigned int);
template void device_conversion<unsigned int,float>(unsigned int*, float*, unsigned int, unsigned int, unsigned int);
template void device_conversion<unsigned char, double>(unsigned char*, double*, unsigned int, unsigned int, unsigned int);





