Hiding dtor from dpcpp

[MrSunshine]

Hi everyone !

I've been struggling for a while on a code, we were on the edge of finishing it but a last problem came by at the last moment.

We have been using a custom type to store a pointer using unified shared memory, and we sucessfully use this custom type inside a kernel on 1D, but when we use more than 1D it creates a temporary type.

And at the end of the kernel scope the temporary type gets deleted. We are using a boolean to prevent the temporary type to avoid freeing memory on the dtor but the compiler reads it anyway.

The easy solution is to free the memory manually outside the kernel, but we just want to tell the compiler not to use the dtor on the CPU side. Is there a preprocessor directive or macro that cancels the segment of code ? 

~MultiArray() {if (m_self_destruct) cl::sycl::free(m_ptr,QUEUE);}

(the segment of code we want dpcpp to avoid reading because m_self_destruct prevents any temporary type to free memory inside the kernel)

#if __SYCL_HOST_ONLY__
~MultiArray() {if (m_self_destruct) cl::sycl::free(m_ptr,QUEUE);}
#endif 

( is that possible ? )

Or any dpcpp options that can accept the error :

error: SYCL kernel cannot call an undefined function without SYCL_EXTERNAL attribute
~MultiArray() {if (m_self_destruct) cl::sycl::free(m_ptr,QUEUE);} 

Or does a pointer using USM free his memory by himself on the dtor even so we do not explicitly call cl::sycl::free ? (the custom type we use is inside another custom class on USM)

MeshGeometry *mesh = new (malloc_shared<MeshGeometry>(sizeof(MeshGeometry),QUEUE)) MeshGeometry();

class MeshGeometry { 
public:
MeshGeometry() { for(int i=0; i<10; i++) m_quads[i] = i; }
private:
  MultiArray<int,10,10> C;
  MultiArray<int,10> m_quads;
};

I wish you all a great day in this difficult time.

[AbhishekD_Intel]

Hi,

Thanks for reaching out to us and for the complete details.

Please give us a small reproducer which will give us more insight into your issue. This will help us to get the idea related to your use-case and will help to provide the exact solution.

Warm Regards,

Abhishek

[MrSunshine]

Hi,

I've made a small package with all files needed for the test.

I usually use cmake to compile and run, but in this example you can compile with dpcpp :

dpcpp simple-sycl-usm-app.cpp -o simple-sycl-usm-app

the error occurs when we call C[i][0] creating a temporary type C[i] that is only a view of the pointer. It is not allocating or freeing any memory.

We would like to keep the dtor to avoid freeing memory manually.

I would like to thank you for the time you dedicate to me.

Warm Regards,

Thomas

[AbhishekD_Intel]

Hi Thomas,

Thanks for the reproducer, it has really helped us in understanding your exact problem.

We tried using the reproducer and we are able to reproduce the same error as you are getting. We are looking into this issue and will get back to you with the details regarding the exact cause of this issue.

Meanwhile, as a workaround don't write anything inside a destructor(keep empty) after the getselfDtor() and try manually freeing the memory.

We will get back to you as soon as we get any updates on the issue.

Warm Regards,

Abhishek

[MrSunshine]

Hi Abhishek,

I would like to thank you for your answer. We will keep an eye on the forum post for any futur update.

We would really like to not freeing memory manually.

Warm Regards,

Thomas.

[Khalik_K_Intel · Accepted solution]

Hello,

Thank you for contacting Intel support.

The problem you are facing is because a temporary type is created when more than 1D custom type is used.

That temporary type invokes dtor, so if you look at the line the error points to - it will be not dtor of 2+D, but dtor of 1D.

The error itself is because you are calling cl::sycl::free from the kernel which is not allowed.

You can allocate and free the memory only from the host.

Usage of temporary type means that it is created on the entry and is deleted at the exit. Therefore, dtor is executed at the kernel exit and you face this error.

Usage of the bool may help you to avoid code being executed in a runtime, when its value is known to a compiler. However, the problem occurs at compile time.

When you take cl::sycl::free from the dtor, you avoid it being called from inside the kernel and that is why the error is not shown.

Unfortunately, Intel DPC++ compiler does not have such option which can accept this error, nor has a preprocessor directive or macro that cancels the segment of code.

The possible options for you, as I can see it, to avoid call of cl::sycl::free and do it manually as you tried already or maybe (if it works for you) avoid usage of 2+D custom types (convert 2D to 1D) which will avoid creation of a temporary type.

Hope this helps.

Regards,

Khalik.

[MrSunshine]

Hi,

Thank you for your answer and for the time you took to analyse my problem.

We will avoid using 2+D arrays inside the device from now. 

Warm Regards,

Thomas. 

[Khalik_K_Intel]

This issue has been resolved and we will no longer respond to this thread. If you require additional assistance from Intel, please start a new thread.

Any further interaction in this thread will be considered community only.