Convert GPU hosted buffer to 3D-image with minimum data copying in OpenCL

In short:

I want to use data stored in GPU buffer as GPU image (3D-texture), if possible, without allocating the memory twice and wasting time by copying the data.

More details:

I have two kernels:

1. First kernel which produce 3D force-field and store it in buffer hosted in GPU memory. I used clCreateBuffer() to allocate this buffer.
2. Second kernel sample force from this force-field at random locations, therefore I use 3D image (allocated by clCreateImage3D()) so that is can use fast gpu-accelerated texture sampling interpolation and border-wrapping. read_imagef() and sampler_t sampler_wrf = CLK_NORMALIZED_COORDS_TRUE | CLK_ADDRESS_MIRRORED_REPEAT | CLK_FILTER_LINEAR;

It would be great if I can convert the buffer produced by kernel (1) into image3D without copying the data and without double-allocation of the GPU memory. The data already are in GPU memory, so I see no point in copying them. Maybe there is a way to just re-interpret the handle cl_mem from buffer to image while keeping the data untouched?

In other words / code:

Now I do it like this:

cl_mem buffer_1 = clCreateBuffer ( ... );
cl_mem image_1  = clCreateImage3D( ... );
clSetKernelArg( krenel_1, ... , &buffer_1 ); 
clEnqueueNDRangeKernel( ..., krenel_1, ... );
clEnqueueCopyBufferToImage( ..., buffer_1, ..., image_1  );
clSetKernelArg( krenel_2, ... , &image_1 ); 
clEnqueueNDRangeKernel( ..., krenel_2, ... );

Is there better / simpler / more *efficient way? By *efficient I mean faster and using less GPU memory.

=======================================

EDIT: I tried to make kernel which writes directly to texture. But does not work. Keep getting error CL_INVALID_KERNEL_ARGS" (-52) when enque the kernell, (clSetKernel Arg() pass without error )

Not sure if this is hardware specific (I have nVidia RTX 2060, driver 515.65.01 on Ubuntu 2022LTS)

cl code (GPU side):

#pragma OPENCL EXTENSION cl_khr_3d_image_writes : enable

__kernel void write_toImg(
  __write_only image3d_t  imgOut
){
    const int ia = get_global_id(0);
    const int ib = get_global_id(1);
    const int ic = get_global_id(2);
    write_imagef( imgOut, (int4){ia,ib,ic,0}, (float){0.0,0.0,0.0,0.0} );
}

C/C++ code (CPU side):

cl_image_format imageFormat={CL_RGBA, CL_FLOAT};
cl_mem itexFF = clCreateImage3D(context, CL_MEM_READ_WRITE, &imageFormat, nImg[0],nImg[1],nImg[2], 0, 0, 0, &err);

size_t global_size[4]{nImg[0],nImg[1],nImg[2],0};
size_t local_size [4]{1,1,1,0};
clSetKernelArg( current_kernel, 0, sizeof(cl_mem), &itexFF) );
clEnqueueNDRangeKernel( commands, current_kernel, 3, NULL, global_size, local_size, 0, NULL, NULL );