May 2020 - Ecp-aurora-sdk-discuss - mailman3.cels.anl.gov

Re: [Ecp-aurora-sdk-discuss] omp/2020.04.30.001 Fortran issue
by Wall, Michael E 19 May '20

19 May '20

Thanks Thomas. Using ifort instead worked for me. Best, Mike ________________________________ From: Applencourt, Thomas <tapplencourt(a)anl.gov> Sent: Tuesday, May 19, 2020 10:50:15 AM To: ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov; Wall, Michael E; Parks, John Subject: [EXTERNAL] Re: omp/2020.04.30.001 Fortran issue Hi Mike, Thanks a lot for the bugs report! We will work with Intel to fix this issue as soon as possible. I assume you used ifx (the new fortran compiler). As a workaround, ifx is ABI compatible with ifort. That means you can compile this file with ifort, and the rest of the code with `ifx`. I know it's a little tedious, to help you can write a little wrapper who tries to do that for you. Something like: ifx $@ if [ $? != 0 ]; then ifort $@ fi Hope this help, Regards, Thomas Thomas Applencourt Leadership Computing Facility Argonne National Laboratory Building 240 Office - 1.D.18 9700 S Cass Av., Lemont Il, 60439 (630) 252-0036 ________________________________ From: Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss-bounces(a)lists.jlse.anl.gov> on behalf of Wall, Michael E via Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov> Sent: Tuesday, May 19, 2020 11:32 To: ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov <ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov> Subject: [Ecp-aurora-sdk-discuss] omp/2020.04.30.001 Fortran issue Hi, I'm encountering a compiler error while building a package that includes Fortran on JLSE. I'm using ifx in the omp/2020.04.30.001 (default) module. The package is https://github.com/lanl/bml Here is the error message: /home/mewall/packages/bml/src/Fortran-interface/bml_allocate_m.F90(99): catastrophic error: **Internal compiler error: internal abort** Please report this error along with the circumstances in which it occurred in a Software Problem Report. Note: File and line given may not be explicit cause of this error. See line 99 of the source code here, with the above Note caveat: https://github.com/lanl/bml/blob/master/src/Fortran-interface/bml_allocate_… Any advice for overcoming this issue would be appreciated. Thank you, Mike

1 0

Re: [Ecp-aurora-sdk-discuss] omp/2020.04.30.001 Fortran issue
by Applencourt, Thomas 19 May '20

19 May '20

Hi Mike, Thanks a lot for the bugs report! We will work with Intel to fix this issue as soon as possible. I assume you used ifx (the new fortran compiler). As a workaround, ifx is ABI compatible with ifort. That means you can compile this file with ifort, and the rest of the code with `ifx`. I know it's a little tedious, to help you can write a little wrapper who tries to do that for you. Something like: ifx $@ if [ $? != 0 ]; then ifort $@ fi Hope this help, Regards, Thomas Thomas Applencourt Leadership Computing Facility Argonne National Laboratory Building 240 Office - 1.D.18 9700 S Cass Av., Lemont Il, 60439 (630) 252-0036 ________________________________ From: Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss-bounces(a)lists.jlse.anl.gov> on behalf of Wall, Michael E via Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov> Sent: Tuesday, May 19, 2020 11:32 To: ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov <ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov> Subject: [Ecp-aurora-sdk-discuss] omp/2020.04.30.001 Fortran issue Hi, I'm encountering a compiler error while building a package that includes Fortran on JLSE. I'm using ifx in the omp/2020.04.30.001 (default) module. The package is https://github.com/lanl/bml Here is the error message: /home/mewall/packages/bml/src/Fortran-interface/bml_allocate_m.F90(99): catastrophic error: **Internal compiler error: internal abort** Please report this error along with the circumstances in which it occurred in a Software Problem Report. Note: File and line given may not be explicit cause of this error. See line 99 of the source code here, with the above Note caveat: https://github.com/lanl/bml/blob/master/src/Fortran-interface/bml_allocate_… Any advice for overcoming this issue would be appreciated. Thank you, Mike

1 0

omp/2020.04.30.001 Fortran issue
by Wall, Michael E 19 May '20

19 May '20

Hi, I'm encountering a compiler error while building a package that includes Fortran on JLSE. I'm using ifx in the omp/2020.04.30.001 (default) module. The package is https://github.com/lanl/bml Here is the error message: /home/mewall/packages/bml/src/Fortran-interface/bml_allocate_m.F90(99): catastrophic error: **Internal compiler error: internal abort** Please report this error along with the circumstances in which it occurred in a Software Problem Report. Note: File and line given may not be explicit cause of this error. See line 99 of the source code here, with the above Note caveat: https://github.com/lanl/bml/blob/master/src/Fortran-interface/bml_allocate_… Any advice for overcoming this issue would be appreciated. Thank you, Mike

1 0

lazy evaluation CRTP expressions in SYCL
by Bryce Allen 13 May '20

13 May '20

I am having issues using CRTP expression templates within SYCL kernels. More details described here: https://github.com/intel/llvm/issues/1687 This is perhaps a less used corner of C++, but I am wondering if anyone else has run into issues with more advanced template features and SYCL kernels? Based on this article I don't think there is anything in the spec which precludes this: https://www.codeplay.com/portal/07-12-19-enabling-polymorphism-in-sycl-usin… Thanks, Bryce

1 0

Re: [Ecp-aurora-sdk-discuss] New version of `iprof`
by Vazquez Mayagoitia, Alvaro 12 May '20

12 May '20

Thanks! iprof is very useful for us, and it works in my laptop too. -Álvaro Vázquez-Mayagoitia, PhD Computational Scientist - R&D Chemistry and Materials Science Argonne National Laboratory https://web.alcf.anl.gov/~vama/ ________________________________ From: Applencourt, Thomas <tapplencourt(a)anl.gov> Sent: Tuesday, May 12, 2020 3:04 PM To: Vazquez Mayagoitia, Alvaro <vama(a)alcf.anl.gov>; ecp-aurora-sdk-discuss(a)jlse.anl.gov <ecp-aurora-sdk-discuss(a)jlse.anl.gov> Subject: Re: New version of `iprof` Hi Alvaro, Good question. We are working with Argonne Layers to OpenSource the new iprof. When they will agree we will update the Github. In the meantime, if you want to have access to the sources on JLSE, they are in `/soft/debuggers/iprof_new/0.2` for the summary script, and for the tracer, you can contact Brice. Hope this is helping, Thomas Thomas Applencourt Leadership Computing Facility Argonne National Laboratory Building 240 Office - 1.D.18 9700 S Cass Av., Lemont Il, 60439 (630) 252-0036 ________________________________ From: Vazquez Mayagoitia, Alvaro <vama(a)alcf.anl.gov> Sent: Tuesday, May 12, 2020 13:07 To: ecp-aurora-sdk-discuss(a)jlse.anl.gov <ecp-aurora-sdk-discuss(a)jlse.anl.gov>; Applencourt, Thomas <tapplencourt(a)anl.gov> Subject: Re: New version of `iprof` Hi Thomas. I wonder if you could update v0.2 here https://github.com/alcf-perfengr/iprof , or could you share the new tool source? thanks -Álvaro Vázquez-Mayagoitia, PhD Computational Scientist - R&D Chemistry and Materials Science Argonne National Laboratory https://web.alcf.anl.gov/~vama/ ________________________________ From: Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss-bounces(a)lists.jlse.anl.gov> on behalf of Applencourt, Thomas via Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov> Sent: Tuesday, May 12, 2020 9:21 AM To: ecp-aurora-sdk-discuss(a)jlse.anl.gov <ecp-aurora-sdk-discuss(a)jlse.anl.gov> Subject: [Ecp-aurora-sdk-discuss] New version of `iprof` Hi, (sorry about the duplicates email) Argonne is pleased to announce a new version of `iprof`. Argonne light way profiling tools for OpenCL and L0. Version 0.1 was OpenCL only and based on Opencl Intercept Layer (https://github.com/intel/opencl-intercept-layer) developed by Intel. It's still the version by default. To use it just type: module load iprof, or module load iprof/0.1 then iprof ./a.out Example of outout: Function Name, Calls, Time (ns), Time (%), Average (ns), Min (ns), Max (ns) clCompileProgram, 6, 1641, 0.00%, 273, 86, 553 clCreateBuffer, 2, 625495, 0.19%, 312747, 7760, 617735 clCreateCommandQueueWithProperties, 1, 1371, 0.00%, 1371, 1371, 1371 clCreateContext, 1, 5927, 0.00%, 5927, 5927, 5927 clCreateKernel, 1, 7700, 0.00%, 7700, 7700, 7700 clCreateProgramWithIL, 6, 60290, 0.02%, 10048, 3033, 22882 clEnqueueMapBuffer, 1, 1290080, 0.39%, 1290080, 1290080, 1290080 clEnqueueNDRangeKernel(hello_world), 1, 252051, 0.08%, 252051, 252051, 252051 clGetContextInfo, 18, 27402, 0.01%, 1522, 118, 20853 clGetDeviceIDs, 8, 4532, 0.00%, 566, 139, 1287 clGetDeviceInfo, 53, 27119, 0.01%, 511, 87, 6727 clGetExtensionFunctionAddressForPlatform, 3, 6222, 0.00%, 2074, 1212, 3347 clGetKernelInfo, 1, 295, 0.00%, 295, 295, 295 clGetPlatformIDs, 2, 115140740, 35.20%, 57570370, 166, 115140574 clGetPlatformInfo, 38, 13171, 0.00%, 346, 103, 992 clLinkProgram, 1, 209642559, 64.09%, 209642559, 209642559, 209642559 clReleaseProgram, 1, 6284, 0.00%, 6284, 6284, 6284 clSetKernelArg, 21, 5076, 0.00%, 241, 123, 835 clSetKernelExecInfo, 3, 961, 0.00%, 320, 127, 692 clWaitForEvents, 1, 1113, 0.00%, 1113, 1113, 1113 Version 0.2 is the version to use our Homemade tracer (developed by @Videau, Brice<mailto:[email protected]>). In short, it does the same thing as v0.1. To use it just type: module load iprof/0.2 One addition is to show you how many byte of memory was explicitly moved by your application. For example: Memory Transfer informations: Name Flags Calls Size (bytes) Size (%) ------------------ --------------------------------------- ------- -------------- ---------- clCreateBuffer CL_MEM_READ_WRITE & CL_MEM_USE_HOST_PTR 2 1033 50.10% clEnqueueMapBuffer CL_MAP_READ 1 1029 49.90% Total 3 2062 Version 0.3 is the first version to support OpenCL and L0 (module load iprof/0.3) . `iprof` will do is maximum to determine if you plan to use OpenCL or L0 backend for OpenMP / SYCL. If you use directly L0 or OpenCL please use `clprof.sh` or `zprof.sh`. The module `iprof_lttng` was been removed. Please use `iprof/0.3`. I'm always looking for feedback or new feature to implements, but If you want as much information possible on your application, Vtune may be a better choice.🙂 Regards, Thomas Thomas Applencourt Leadership Computing Facility Argonne National Laboratory Building 240 Office - 1.D.18 9700 S Cass Av., Lemont Il, 60439 (630) 252-0036

1 0

Re: [Ecp-aurora-sdk-discuss] New version of `iprof`
by Applencourt, Thomas 12 May '20

12 May '20

Hi Alvaro, Good question. We are working with Argonne Layers to OpenSource the new iprof. When they will agree we will update the Github. In the meantime, if you want to have access to the sources on JLSE, they are in `/soft/debuggers/iprof_new/0.2` for the summary script, and for the tracer, you can contact Brice. Hope this is helping, Thomas Thomas Applencourt Leadership Computing Facility Argonne National Laboratory Building 240 Office - 1.D.18 9700 S Cass Av., Lemont Il, 60439 (630) 252-0036 ________________________________ From: Vazquez Mayagoitia, Alvaro <vama(a)alcf.anl.gov> Sent: Tuesday, May 12, 2020 13:07 To: ecp-aurora-sdk-discuss(a)jlse.anl.gov <ecp-aurora-sdk-discuss(a)jlse.anl.gov>; Applencourt, Thomas <tapplencourt(a)anl.gov> Subject: Re: New version of `iprof` Hi Thomas. I wonder if you could update v0.2 here https://github.com/alcf-perfengr/iprof , or could you share the new tool source? thanks -Álvaro Vázquez-Mayagoitia, PhD Computational Scientist - R&D Chemistry and Materials Science Argonne National Laboratory https://web.alcf.anl.gov/~vama/ ________________________________ From: Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss-bounces(a)lists.jlse.anl.gov> on behalf of Applencourt, Thomas via Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov> Sent: Tuesday, May 12, 2020 9:21 AM To: ecp-aurora-sdk-discuss(a)jlse.anl.gov <ecp-aurora-sdk-discuss(a)jlse.anl.gov> Subject: [Ecp-aurora-sdk-discuss] New version of `iprof` Hi, (sorry about the duplicates email) Argonne is pleased to announce a new version of `iprof`. Argonne light way profiling tools for OpenCL and L0. Version 0.1 was OpenCL only and based on Opencl Intercept Layer (https://github.com/intel/opencl-intercept-layer) developed by Intel. It's still the version by default. To use it just type: module load iprof, or module load iprof/0.1 then iprof ./a.out Example of outout: Function Name, Calls, Time (ns), Time (%), Average (ns), Min (ns), Max (ns) clCompileProgram, 6, 1641, 0.00%, 273, 86, 553 clCreateBuffer, 2, 625495, 0.19%, 312747, 7760, 617735 clCreateCommandQueueWithProperties, 1, 1371, 0.00%, 1371, 1371, 1371 clCreateContext, 1, 5927, 0.00%, 5927, 5927, 5927 clCreateKernel, 1, 7700, 0.00%, 7700, 7700, 7700 clCreateProgramWithIL, 6, 60290, 0.02%, 10048, 3033, 22882 clEnqueueMapBuffer, 1, 1290080, 0.39%, 1290080, 1290080, 1290080 clEnqueueNDRangeKernel(hello_world), 1, 252051, 0.08%, 252051, 252051, 252051 clGetContextInfo, 18, 27402, 0.01%, 1522, 118, 20853 clGetDeviceIDs, 8, 4532, 0.00%, 566, 139, 1287 clGetDeviceInfo, 53, 27119, 0.01%, 511, 87, 6727 clGetExtensionFunctionAddressForPlatform, 3, 6222, 0.00%, 2074, 1212, 3347 clGetKernelInfo, 1, 295, 0.00%, 295, 295, 295 clGetPlatformIDs, 2, 115140740, 35.20%, 57570370, 166, 115140574 clGetPlatformInfo, 38, 13171, 0.00%, 346, 103, 992 clLinkProgram, 1, 209642559, 64.09%, 209642559, 209642559, 209642559 clReleaseProgram, 1, 6284, 0.00%, 6284, 6284, 6284 clSetKernelArg, 21, 5076, 0.00%, 241, 123, 835 clSetKernelExecInfo, 3, 961, 0.00%, 320, 127, 692 clWaitForEvents, 1, 1113, 0.00%, 1113, 1113, 1113 Version 0.2 is the version to use our Homemade tracer (developed by @Videau, Brice<mailto:[email protected]>). In short, it does the same thing as v0.1. To use it just type: module load iprof/0.2 One addition is to show you how many byte of memory was explicitly moved by your application. For example: Memory Transfer informations: Name Flags Calls Size (bytes) Size (%) ------------------ --------------------------------------- ------- -------------- ---------- clCreateBuffer CL_MEM_READ_WRITE & CL_MEM_USE_HOST_PTR 2 1033 50.10% clEnqueueMapBuffer CL_MAP_READ 1 1029 49.90% Total 3 2062 Version 0.3 is the first version to support OpenCL and L0 (module load iprof/0.3) . `iprof` will do is maximum to determine if you plan to use OpenCL or L0 backend for OpenMP / SYCL. If you use directly L0 or OpenCL please use `clprof.sh` or `zprof.sh`. The module `iprof_lttng` was been removed. Please use `iprof/0.3`. I'm always looking for feedback or new feature to implements, but If you want as much information possible on your application, Vtune may be a better choice.🙂 Regards, Thomas Thomas Applencourt Leadership Computing Facility Argonne National Laboratory Building 240 Office - 1.D.18 9700 S Cass Av., Lemont Il, 60439 (630) 252-0036

1 0

Re: [Ecp-aurora-sdk-discuss] New version of `iprof`
by Vazquez Mayagoitia, Alvaro 12 May '20

12 May '20

Hi Thomas. I wonder if you could update v0.2 here https://github.com/alcf-perfengr/iprof , or could you share the new tool source? thanks -Álvaro Vázquez-Mayagoitia, PhD Computational Scientist - R&D Chemistry and Materials Science Argonne National Laboratory https://web.alcf.anl.gov/~vama/ ________________________________ From: Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss-bounces(a)lists.jlse.anl.gov> on behalf of Applencourt, Thomas via Ecp-aurora-sdk-discuss <ecp-aurora-sdk-discuss(a)lists.jlse.anl.gov> Sent: Tuesday, May 12, 2020 9:21 AM To: ecp-aurora-sdk-discuss(a)jlse.anl.gov <ecp-aurora-sdk-discuss(a)jlse.anl.gov> Subject: [Ecp-aurora-sdk-discuss] New version of `iprof` Hi, (sorry about the duplicates email) Argonne is pleased to announce a new version of `iprof`. Argonne light way profiling tools for OpenCL and L0. Version 0.1 was OpenCL only and based on Opencl Intercept Layer (https://github.com/intel/opencl-intercept-layer) developed by Intel. It's still the version by default. To use it just type: module load iprof, or module load iprof/0.1 then iprof ./a.out Example of outout: Function Name, Calls, Time (ns), Time (%), Average (ns), Min (ns), Max (ns) clCompileProgram, 6, 1641, 0.00%, 273, 86, 553 clCreateBuffer, 2, 625495, 0.19%, 312747, 7760, 617735 clCreateCommandQueueWithProperties, 1, 1371, 0.00%, 1371, 1371, 1371 clCreateContext, 1, 5927, 0.00%, 5927, 5927, 5927 clCreateKernel, 1, 7700, 0.00%, 7700, 7700, 7700 clCreateProgramWithIL, 6, 60290, 0.02%, 10048, 3033, 22882 clEnqueueMapBuffer, 1, 1290080, 0.39%, 1290080, 1290080, 1290080 clEnqueueNDRangeKernel(hello_world), 1, 252051, 0.08%, 252051, 252051, 252051 clGetContextInfo, 18, 27402, 0.01%, 1522, 118, 20853 clGetDeviceIDs, 8, 4532, 0.00%, 566, 139, 1287 clGetDeviceInfo, 53, 27119, 0.01%, 511, 87, 6727 clGetExtensionFunctionAddressForPlatform, 3, 6222, 0.00%, 2074, 1212, 3347 clGetKernelInfo, 1, 295, 0.00%, 295, 295, 295 clGetPlatformIDs, 2, 115140740, 35.20%, 57570370, 166, 115140574 clGetPlatformInfo, 38, 13171, 0.00%, 346, 103, 992 clLinkProgram, 1, 209642559, 64.09%, 209642559, 209642559, 209642559 clReleaseProgram, 1, 6284, 0.00%, 6284, 6284, 6284 clSetKernelArg, 21, 5076, 0.00%, 241, 123, 835 clSetKernelExecInfo, 3, 961, 0.00%, 320, 127, 692 clWaitForEvents, 1, 1113, 0.00%, 1113, 1113, 1113 Version 0.2 is the version to use our Homemade tracer (developed by @Videau, Brice<mailto:[email protected]>). In short, it does the same thing as v0.1. To use it just type: module load iprof/0.2 One addition is to show you how many byte of memory was explicitly moved by your application. For example: Memory Transfer informations: Name Flags Calls Size (bytes) Size (%) ------------------ --------------------------------------- ------- -------------- ---------- clCreateBuffer CL_MEM_READ_WRITE & CL_MEM_USE_HOST_PTR 2 1033 50.10% clEnqueueMapBuffer CL_MAP_READ 1 1029 49.90% Total 3 2062 Version 0.3 is the first version to support OpenCL and L0 (module load iprof/0.3) . `iprof` will do is maximum to determine if you plan to use OpenCL or L0 backend for OpenMP / SYCL. If you use directly L0 or OpenCL please use `clprof.sh` or `zprof.sh`. The module `iprof_lttng` was been removed. Please use `iprof/0.3`. I'm always looking for feedback or new feature to implements, but If you want as much information possible on your application, Vtune may be a better choice.🙂 Regards, Thomas Thomas Applencourt Leadership Computing Facility Argonne National Laboratory Building 240 Office - 1.D.18 9700 S Cass Av., Lemont Il, 60439 (630) 252-0036

1 0

New version of `iprof`
by Applencourt, Thomas 12 May '20

12 May '20

Hi, (sorry about the duplicates email) Argonne is pleased to announce a new version of `iprof`. Argonne light way profiling tools for OpenCL and L0. Version 0.1 was OpenCL only and based on Opencl Intercept Layer (https://github.com/intel/opencl-intercept-layer) developed by Intel. It's still the version by default. To use it just type: module load iprof, or module load iprof/0.1 then iprof ./a.out Example of outout: Function Name, Calls, Time (ns), Time (%), Average (ns), Min (ns), Max (ns) clCompileProgram, 6, 1641, 0.00%, 273, 86, 553 clCreateBuffer, 2, 625495, 0.19%, 312747, 7760, 617735 clCreateCommandQueueWithProperties, 1, 1371, 0.00%, 1371, 1371, 1371 clCreateContext, 1, 5927, 0.00%, 5927, 5927, 5927 clCreateKernel, 1, 7700, 0.00%, 7700, 7700, 7700 clCreateProgramWithIL, 6, 60290, 0.02%, 10048, 3033, 22882 clEnqueueMapBuffer, 1, 1290080, 0.39%, 1290080, 1290080, 1290080 clEnqueueNDRangeKernel(hello_world), 1, 252051, 0.08%, 252051, 252051, 252051 clGetContextInfo, 18, 27402, 0.01%, 1522, 118, 20853 clGetDeviceIDs, 8, 4532, 0.00%, 566, 139, 1287 clGetDeviceInfo, 53, 27119, 0.01%, 511, 87, 6727 clGetExtensionFunctionAddressForPlatform, 3, 6222, 0.00%, 2074, 1212, 3347 clGetKernelInfo, 1, 295, 0.00%, 295, 295, 295 clGetPlatformIDs, 2, 115140740, 35.20%, 57570370, 166, 115140574 clGetPlatformInfo, 38, 13171, 0.00%, 346, 103, 992 clLinkProgram, 1, 209642559, 64.09%, 209642559, 209642559, 209642559 clReleaseProgram, 1, 6284, 0.00%, 6284, 6284, 6284 clSetKernelArg, 21, 5076, 0.00%, 241, 123, 835 clSetKernelExecInfo, 3, 961, 0.00%, 320, 127, 692 clWaitForEvents, 1, 1113, 0.00%, 1113, 1113, 1113 Version 0.2 is the version to use our Homemade tracer (developed by @Videau, Brice<mailto:[email protected]>). In short, it does the same thing as v0.1. To use it just type: module load iprof/0.2 One addition is to show you how many byte of memory was explicitly moved by your application. For example: Memory Transfer informations: Name Flags Calls Size (bytes) Size (%) ------------------ --------------------------------------- ------- -------------- ---------- clCreateBuffer CL_MEM_READ_WRITE & CL_MEM_USE_HOST_PTR 2 1033 50.10% clEnqueueMapBuffer CL_MAP_READ 1 1029 49.90% Total 3 2062 Version 0.3 is the first version to support OpenCL and L0 (module load iprof/0.3) . `iprof` will do is maximum to determine if you plan to use OpenCL or L0 backend for OpenMP / SYCL. If you use directly L0 or OpenCL please use `clprof.sh` or `zprof.sh`. The module `iprof_lttng` was been removed. Please use `iprof/0.3`. I'm always looking for feedback or new feature to implements, but If you want as much information possible on your application, Vtune may be a better choice.🙂 Regards, Thomas Thomas Applencourt Leadership Computing Facility Argonne National Laboratory Building 240 Office - 1.D.18 9700 S Cass Av., Lemont Il, 60439 (630) 252-0036

1 0

Re: [Ecp-aurora-sdk-discuss] USM allocations and access
by Vincent R. Pascuzzi 07 May '20

07 May '20

Hi, James. > It's not guaranteed to work, and on systems with a discrete GPU, will likely not. I can build and run the same code, modulo the `cout`s, using the Intel llvm CUDA support: clang++ -g -O2 -fsycl -std=c++17 -fsycl-targets=nvptx64-nvidia-cuda-sycldevice -Wno-unknown-cuda-version -o mwe_malloc_device mwe_malloc_device.cc I get a segfault when trying to access the device memory outside a kernel — i.e. using `cout`. If I replace the `cout` calls with `cl::sycl::stream`s, it works as expected on the CUDA device: ``` … // Outputs 42 queue .submit([&](cl::sycl::handler& cgh) { cl::sycl::stream out(1024, 256, cgh); cgh.single_task<class print1>( [=] { out << deviceArray[10] << cl::sycl::endl; }); }) .wait_and_throw(); queue.submit([&](cl::sycl::handler& h) { h.parallel_for<class foo>( cl::sycl::range<1>{42}, // lambda-capture so we get the actual device memory [=](cl::sycl::id<1> ID) { int i = ID[0]; deviceArray[i]++; }); }); queue.wait(); // Outputs 43 queue .submit([&](cl::sycl::handler& cgh) { cl::sycl::stream out(1024, 256, cgh); cgh.single_task<class print2>( [=] { out << deviceArray[10] << cl::sycl::endl; }); }) .wait_and_throw(); return 0; ``` So indeed USM works fine in this simple example. > Device pointers happen to work on systems with integrated Intel GPUs since there is no actual device memory - it's all shared with the host CPU. This one sentence answers my questions — thanks! Vince

1 0

USM allocations and access
by Vincent R. Pascuzzi 07 May '20

07 May '20

Hi, (Sorry if this is the wrong list.) I am using `malloc_device` to allocate device-side (Iris iGPU) memory through USM. An example of this is, ``` int hostArray[42]; int* deviceArray = (int*)malloc_device(42 * sizeof(int), dev, ctx_); for (int i = 0; i < 42; i++) hostArray[i] = 42; queue_.submit([&](cl::sycl::handler& h) { // copy hostArray to deviceArray h.memcpy(deviceArray, &hostArray[0], 42 * sizeof(int)); }); queue_.wait(); ``` Now, what’s odd to me is that I can access `deviceArray` outside a kernel, like this: ``` std::cout << "deviceArray[10] = " << deviceArray[10] << std::endl; ``` Because of “unified addressing” [1], and since all USM allocations are done “on the host” [2], I assumed that I was outputting the `hostArray` memory address. Fine. But then I modified the value of `deviceArray[10]` inside a kernel: ``` queue_.submit([&](cl::sycl::handler& h) { h.parallel_for<class foo>(cl::sycl::range<1>{42}, [=](cl::sycl::id<1> ID) { int i = ID[0]; deviceArray[i]++; }); }); queue_.wait(); ``` and I was still able to access `deviceArray` outside a kernel in the same way as above, and modified value was printed. I can understand this working fine if I was using the “host device” but can’t wrap my head around accessing “GPU device” memory outside a kernel, especially if it has been modified inside a kernel. A complete MWE is at the bottom of this email. I suggest running on an Iris node. Can someone explain this to me? Why/how is it possible to access memory allocated with `malloc_device` outside a kernel? If it’s because it’s allocated by the host — and is duplicated on the host — why when modifications are made inside a kernel does it affect the host memory? `malloc_shared` — wherein data is migrated back and forth between the host and device — would be a different story. Thanks, Vince [1] https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/USM/USM.adoc [2] https://www.colfax-intl.com/downloads/oneAPI_module04_DPCplusplusFundamenta… (slide 23) // Setup environment and build module load dpcpp/2020.04.30.001 [dpcpp | clang++] -fsycl -o mwe_malloc_device mwe_malloc_device.cc ``` // // mwe_malloc_device.cc // #include <CL/sycl.hpp> #include <iostream> int main() { // Catch asynchronous exceptions auto exception_handler = [](cl::sycl::exception_list exceptions) { for (std::exception_ptr const& e : exceptions) { try { std::rethrow_exception(e); } catch (cl::sycl::exception const& e) { std::cout << "Caught asynchronous SYCL exception during generation:\n" << e.what() << std::endl; } } }; // Initialize device, queue and context cl::sycl::device dev; #ifdef USE_PI_CUDA CUDASelector cuda_selector; try { dev = cl::sycl::device(cuda_selector); } catch (...) { } #elif USE_SYCL_CPU dev = cl::sycl::device(cl::sycl::cpu_selector()); #elif USE_SYCL_GPU dev = cl::sycl::device(cl::sycl::gpu_selector()); #else dev = cl::sycl::device(cl::sycl::default_selector()); #endif cl::sycl::queue queue = cl::sycl::queue(dev, exception_handler); cl::sycl::context ctx = queue.get_context(); // Name of the device to run on std::string dev_name = queue.get_device().get_info<cl::sycl::info::device::name>(); std::cout << "Using device \"" << dev_name << "\"" << std::endl; // Ensure device can handle USM device allocations. if (!queue.get_device() .get_info<cl::sycl::info::device::usm_device_allocations>()) { std::cout << "ERROR :: device \"" << dev_name << "\" does not support usm_device_allocations!" << std::endl; return 1; } int hostArray[42]; int* deviceArray = (int*)malloc_device(42 * sizeof(int), dev, ctx); for (int i = 0; i < 42; i++) hostArray[i] = 42; queue .submit([&](cl::sycl::handler& h) { // copy hostArray to deviceArray h.memcpy(deviceArray, &hostArray[0], 42 * sizeof(int)); }) .wait(); std::cout << "[Before mod] deviceArray[10] = " << deviceArray[10] << std::endl; queue.submit([&](cl::sycl::handler& h) { h.parallel_for<class foo>(cl::sycl::range<1>{42}, // lambda-capture `deviceArray` so we get the actual device memory [=, dev_arr = deviceArray](cl::sycl::id<1> ID) { int i = ID[0]; dev_arr[i]++; }); }); queue.wait(); std::cout << "[After mod] deviceArray[10] = " << deviceArray[10] << std::endl; return 0; } ```

1 0