Merge pull request #2763 from CharlesCongdon/cc_tutupdate20260_feedback_updates

Address concerns raised with pull request #2762

Author: chuanyuf

Tools/ApplicationDebugger/guided_matrix_mult_BadBuffers/README.md

@@ -543,7 +543,7 @@ In `b1_matrix_mul_null_usm.cpp` a bad (in this case, null) pointer that is suppo
 
 #### Debugging the Problem
 
-Why did we try with multiple backends?   If one had shown correct or incorrect results, and one had crashed, we might be facing a race condition that only occasionally manifests when something goes terribly wrong.  Or one of the backbends might have a bug while the others do not.  But here all three crash, so it's likely the program is doing something illegal to memory.  The host CPU is a particularly good place to test for illegal memory accesses, because the CPU never allows pointers with an address within a few kilobytes of address `0x0`, while this may be legally allocated memory on the GPU.
+Why did we try with multiple backends?   If one had shown correct or incorrect results, and one had crashed, we might be facing a race condition that only occasionally manifests when something goes terribly wrong.  Or one of the backends might have a bug while the others do not.  But here all three crash, so it's likely the program is doing something illegal to memory.  The host CPU is a particularly good place to test for illegal memory accesses, because the CPU never allows pointers with an address within a few kilobytes of address `0x0`, while this may be legally allocated memory on the GPU.
 
 Another reason to try different backends is that debugging support may differ between different GPU drivers and/or different GPU models.  Debugging the program using the OpenCL™ CPU driver gets around these issues.
 

Tools/ApplicationDebugger/guided_matrix_mult_BadBuffers/src/b1_matrix_mul_null_usm.cpp

@@ -133,7 +133,11 @@ int main() {
 
     q.wait();
 
-    sycl::free(dev_a, q);
+    sycl::free(dev_a, q);   // For the purposes of this demo, since we lost the original pointer value
+                            // and will crash before this point, this free is effectively a no-op and a
+                            // bug in its own right (we should be checking pointers best we can before
+                            // trying to free them).   
+                            // All USM allocations should be released after use.
     sycl::free(dev_b, q);
     sycl::free(dev_c, q);
   }

Tools/ApplicationDebugger/guided_matrix_mult_InvalidContexts/README.md

@@ -174,7 +174,7 @@ To complete the steps in the following section, you must download the [Unified T
    [opencl:gpu][opencl:2] Intel(R) OpenCL Graphics, Intel(R) Data Center GPU Max 1550 OpenCL 3.0 NEO  [25.18.33578]
    [opencl:cpu][opencl:3] Intel(R) OpenCL, Intel(R) Xeon(R) Platinum 8360Y CPU @ 2.40GHz OpenCL 3.0 (Build 0) [2023.16.7.0.21_160000]
    ```
-   > **Note:** If you have only one `[level_zero:gpu]` device listed, or the order is different from the above, the the main example below may not work.   Try to follow through anyway, and then try the bonus sample at the end of this document, which should work no matter what system configuration.
+   > **Note:** If you have only one `[level_zero:gpu]` device listed, or the order is different from the above, the main example below may not work.   Try to follow through anyway, and then try the bonus sample at the end of this document, which should work no matter what system configuration.
 
 3. SYCL applications use the Level Zero runtime by default with an Intel GPU.  But what happens if you use OpenCL™ runtime to run `1_matrix_mul_invalid_contexts` on the GPU?
 
@@ -461,7 +461,7 @@ In case we need view code running on the GPU, we need to enable GPU debugging.
 
 ### Identify the Problem without Code Inspection
 
-You need to build the [Unified Tracing and Profiling Tool](#getting-the-tracing-and-profiling-tool) before completing this section. Once you have built the utility, you can start it before your program (similar to using GBD).
+You need to build the [Unified Tracing and Profiling Tool](#getting-the-tracing-and-profiling-tool) before completing this section. Once you have built the utility, you can start it before your program (similar to using GDB).
 
 One of the things that the Unified Tracing and Profiling utility can help us see
 is every low-level API call made to OpenCL™ or Level Zero. We will use it to attempt to match the source to the events.
@@ -633,7 +633,7 @@ be entirely re-written, which is beyond the scope of this tutorial.
    101     queue q2(devicecontext, selected_device);
    102     float * dev_c = sycl::malloc_device<float>(M*P, q2);
    ```
-   As is hopefully obvious from the previous example, the problem is that we are trying to free memory allocated in SYCL queue `q2` that has a different device context fron SYCL queue `q`, even though under the covers they point to the same hardware device.
+   As is hopefully obvious from the previous example, the problem is that we are trying to free memory allocated in SYCL queue `q2` that has a different device context from SYCL queue `q`, even though under the covers they point to the same hardware device.
 
 6. While we are at it, what does the debugger show us?
    ```

Tools/ApplicationDebugger/guided_matrix_mult_InvalidContexts/src/1_matrix_mul_invalid_contexts.cpp

@@ -79,7 +79,7 @@ int main() {
     }
 
     // Be very specific about the device to use.
-    queue q(devices[0]);
+    queue q(devices[0], propList);
 
     cout << "Computing" << "\n";
     cout << "Device: " << q.get_device().get_info<info::device::name>() << "\n";
@@ -95,7 +95,7 @@ int main() {
 #ifdef BAD_FREE
     device selected_device = devices[0];
 #else
-    device selected_device = devices[1];
+    device selected_device = devices.size() > 1 ? devices[1] : devices[0];
 #endif
     context devicecontext(selected_device);
     queue q2(devicecontext, selected_device);

Tools/ApplicationDebugger/guided_matrix_mult_InvalidContexts/src/2_matrix_mul.cpp

@@ -78,7 +78,7 @@ int main() {
                 << "] " << backend << ",  " << vendor <<  std::endl;
     }
 
-    queue q(default_selector_v);
+    queue q(default_selector_v, propList);
 
     cout << "Computing" << "\n";
     cout << "Device: " << q.get_device().get_info<info::device::name>() << "\n";

Tools/ApplicationDebugger/guided_matrix_mult_RaceCondition/README.md

@@ -336,7 +336,7 @@ Unfortunately not; pretty much the same thing happens - they both produce incorr
       export PATH=$PATH:/opt/intel/oneapi:$DRIVERLOC
    ```
 
-Similarly, we an force the program to run on the CPU, which sometimes can avoid problems in your code that are specific to offloading to the GPU.
+Similarly, we can force the program to run on the CPU, which sometimes can avoid problems in your code that are specific to offloading to the GPU.
 ```
 ONEAPI_DEVICE_SELECTOR=*:cpu ./1_matrix_mul_race_condition
 ```

Tools/ApplicationDebugger/guided_matrix_mult_SLMSize/README.md

@@ -287,7 +287,7 @@ Among other things, the Tracing and Profiling utility can print every low-level
    >>>> [672116802588167] zeCommandListAppendLaunchKernel: hCommandList = 0x3d56268 hKernel = 0x2689658 (_ZTSZZ4mainENKUlRN4sycl3_V17handlerEE_clES2_EUlNS0_7nd_itemILi1EEEE_) pLaunchFuncArgs = 0x7fffb216c840 {16385, 1, 1} hSignalEvent = 0x3b585a8 numWaitEvents = 0x0 phWaitEvents = 0x0
    ```
 
-   At like 106 we used the form of `parallel_for` that takes the `nd_range`, which specifies the global iteration range (163850) and the local work-group size (10) like so:  `nd_range<1>{{163850}, {10}}`. The first line above shows the workgroup size (`groupSizeX = 0xa groupSizeY = 0x1 groupSizeZ = 0x1`), and the second shows how many total workgroups will be needed to process the global iteration range (`{16385, 1, 1}`).
+   At line 106 we used the form of `parallel_for` that takes the `nd_range`, which specifies the global iteration range (163850) and the local work-group size (10) like so:  `nd_range<1>{{163850}, {10}}`. The first line above shows the workgroup size (`groupSizeX = 0xa groupSizeY = 0x1 groupSizeZ = 0x1`), and the second shows how many total workgroups will be needed to process the global iteration range (`{16385, 1, 1}`).
 
 ### Determine Device Limits