done bringing in device code from operations.cu, otherwise linking doesn't find __device__ functions