CUDA, prevent destructor call when passing struct to kernel while keeping the copy constructor usable

I have the following struct:

struct data {
    data(const data& other) {
        other.is_original = false;

    ~data() {
        if(is_original) {
            // Free host allocated data, that we want to keep after the kernel finishes 

And the following kernel:

__global__ void test_kernel(data d) {
    // ...

If I call the kernel now, and pass a struct into it, the copy constructor gets invoked, and, on the new copy of the original struct, the is_original flag is toggled to false. This flag is then used when the destructor gets called to determine, whether we actually want to destroy our data, or not.

How can I modify my code (or completely rework it), so that the copy constructor can be used as a regular copy constructor, while still being able to differentiate between destructor calls. Note that I’m just trying to get around the destructor being called because we’re passing by value.

Also note that I’m aware that this is a frequent question on the forums, it’s just that by searching myself I was unable to find a solution that got further than me.