 	gdf_size_type num_chars_bitmask = ( ( input->size +( GDF_VALID_BITSIZE - 1)) / GDF_VALID_BITSIZE );
 	thrust::copy(thrust::cuda::par.on(stream), input->valid, input->valid + num_chars_bitmask, output->valid); // copy over valid bitmask
 
-	switch ( input->dtype ) {
-	    case    GDF_DATE64:
-	    	thrust::device_ptr<int64_t> input_ptr((int64_t *) input->data);
-	    	thrust::device_ptr<int16_t> output_ptr((int16_t *) output->data);
-	    	gdf_extract_datetime_second_date64_op op;
-	    	thrust::transform(thrust::cuda::par.on(stream), thrust::detail::make_normal_iterator(input_ptr),
-	    			thrust::detail::make_normal_iterator(input_ptr) + input->size, hrust::detail::make_normal_iterator(output_ptr), op);
-
-	    default: return GDF_UNSUPPORTED_DTYPE;
-	}
-
-	cudaStreamSynchronize(stream);
-	cudaStreamDestroy(stream);
-
-	return GDF_SUCCESS;
-}
-
-
-
+	if ( input->dtype == GDF_DATE64 ) {
+		thrust::device_ptr<int64_t> input_ptr((int64_t *) input->data);
+		thrust::device_ptr<int16_t> output_ptr((int16_t *) output->data);
+		gdf_extract_datetime_second_date64_op op;
+		thrust::transform(thrust::cuda::par.on(stream), thrust::detail::make_normal_iterator(input_ptr),
+				thrust::detail::make_normal_iterator(input_ptr) + input->size, thrust::detail::make_normal_iterator(output_ptr), op);
+
+	} else {
+		return GDF_UNSUPPORTED_DTYPE;
+	}
+
+	cudaStreamSynchronize(stream);
+	cudaStreamDestroy(stream);
+
+	return GDF_SUCCESS;
+}
+
+
+

	gdf_size_type num_chars_bitmask = ( ( input->size +( GDF_VALID_BITSIZE - 1)) / GDF_VALID_BITSIZE );
	thrust::copy(thrust::cuda::par.on(stream), input->valid, input->valid + num_chars_bitmask, output->valid); // copy over valid bitmask

	switch ( input->dtype ) {
	    case    GDF_DATE64:
	    	thrust::device_ptr<int64_t> input_ptr((int64_t *) input->data);
	    	thrust::device_ptr<int16_t> output_ptr((int16_t *) output->data);
	    	gdf_extract_datetime_year_date64_op op;
	    	thrust::transform(thrust::cuda::par.on(stream), thrust::detail::make_normal_iterator(input_ptr),
	    			thrust::detail::make_normal_iterator(input_ptr) + input->size, hrust::detail::make_normal_iterator(output_ptr), op);

	    case   GDF_DATE32:
	    	thrust::device_ptr<int32_t> input_ptr((int32_t *) input->data);
	    	thrust::device_ptr<int16_t> output_ptr((int16_t *) output->data);
	    	gdf_extract_datetime_year_date32_op op;
	    	thrust::transform(thrust::cuda::par.on(stream), thrust::detail::make_normal_iterator(input_ptr),
	    			thrust::detail::make_normal_iterator(input_ptr) + input->size, hrust::detail::make_normal_iterator(output_ptr), op);
	    default: return GDF_UNSUPPORTED_DTYPE;
	}

	cudaStreamSynchronize(stream);


1528935590000
'2018-06-14 00:19:50'
 */

struct gdf_extract_datetime_hour_date64_op : public thrust::unary_function<int64_t, int16_t>
{
