How to create a custom forward iterator for thrust?

Accelerated Computing GPU-Accelerated Libraries
1

I have a float4 type from a library and an array of these float4s in GPU memory. I want to use thrust to set the last element of each float4 in the array to a certain value. I figured I could use thrust:fill and a custom iterator like shown, but I am getting an error (also shown). What am I doing wrong, help is greatly appreciated.

error: invalid application of ‘sizeof’ to incomplete type ‘thrust::detail::STATIC_ASSERTION_FAILURE’
sizeof(::thrust::detail::STATIC_ASSERTION_FAILURE< (bool)( B ) >)>\

class Float4Iterator { //I implement these functions appropriately 
public:
    __host__ __device__ Float4Iterator();
    __host__ __device__ Float4Iterator(sl::float4 *data, int dim);
    
    __host__ __device__ float & operator*();
    __host__ __device__ Float4Iterator& operator++();
    __host__ __device__ bool operator!=(Float4Iterator rhs);

    __host__ __device__ Float4Iterator& operator++(int dummy);
    __host__ __device__ bool operator==(Float4Iterator rhs);


private:
   sl::float4* elem;
   int dim;
};

Float4Iterator start(pc_f4.data, 3);
Float4Iterator stop();
thrust::fill(thrust::device, start, stop, 0.0);