LI82341: DEBUGABILITY OF GPU CODE WHEN USING NO-OPT FLAGS.

APAR status

• Closed as program error.

Error description

• When attempting to print values in an OMP CUDA section, the
  values are reported as 'optimized out' when the source code is
  compiled with no-opt flags.
  
  
  
     === EXAMPLE ===
  
  SOURCE:
  
  double table[size];
  ...
  #pragma omp target teams distribute parallel for
     map(from:table[0:256])
      for(int n=0; n<size; ++n) {
        table[n] = val(n);
  
  ====================================
  
  GDB DEBUGGER:
  
  [Switching focus to CUDA kernel 0, grid 2, block (0,0,0), thread
  (96,0,0), device 0, sm 0, warp 0, lane 0]
  0x0000000010d3ef10 in __xl_main_l16_OL_1<<<(2,1,1),(128,1,1)>>>
  (
      table=<optimized out>)
      at test.c:23
  23             for (counter = 0; counter < 10000000; counter++)
  {
  (cuda-gdb) print table
  $1 = <optimized out>
  

Local fix

• N/A
  

Problem summary

• PROBLEM DESCRIPTION:
  Values in an OMP CUDA section do not print out during
  debugging.
  
  USERS AFFECTED:
  Users debugging GPU CUDA code with GDB.
  

Problem conclusion

• With this update, renaming the kernel parameter name to match
  the name of the original variable will allow debugging of the
  target variable.
  

Temporary fix

Comments

APAR Information

• APAR is sysrouted FROM one or more of the following:

• APAR is sysrouted TO one or more of the following:

Fix information

• Fixed component name

  XL C/C++ LINUX

• Fixed component ID

  5725C7310

Applicable component levels