Adding MonteCarloMultiGPU in Jupyter training (#2357)

* adding MonteCarloMultiGpu in training

* Added 05_OceanFFT and 06_SimpleCudaGraph change

---------

Co-authored-by: AvijitBag07 <[email protected]>

Author: AvijitBag07

DirectProgramming/C++SYCL/Jupyter/cuda-to-sycl-migration-training/05_SYCL_Migration_OceanFFT/05_SYCL_Migration_OceanFFT.ipynb

@@ -175,10 +175,31 @@
     "\n",
     "To compile the migrated SYCL code we can use the following command:\n",
     "```\n",
-    "icpx -fsycl -fsycl-targets=intel_gpu_pvc -I ../../../Common -I ../../../include *.cpp -qmkl \n",
+    "icpx -fsycl -fsycl-targets=intel_gpu_pvc -I ../../../Common -I ../../../include *.cpp\n",
     "```\n",
     "\n",
     "There may be compile errors based on whether all of the CUDA code was migrated to SYCL or not. The migrated code may also include comments with warning messages, which could help make it easier to fix the errors. These errors have to be manually fixed to get the code to compile.\n",
+    "\n",
+    "#### Build and Run\n",
+    "Select the cell below and click run ▶ to compile and execute the code (expect to see errors):"
+   ]
+  },
+  {
+   "cell_type": "code",
+   "execution_count": null,
+   "id": "0928ee62-95f4-4cdc-85eb-7cc536af5c49",
+   "metadata": {},
+   "outputs": [],
+   "source": [
+    "! ./q.sh run_dpct_output.sh"
+   ]
+  },
+  {
+   "cell_type": "markdown",
+   "id": "641af418-8989-4523-9e21-5bc62f3ac5f3",
+   "metadata": {},
+   "source": [
+
     "\n",
     "\n",
     "### Compilation Error in migrated SYCL code\n",

DirectProgramming/C++SYCL/Jupyter/cuda-to-sycl-migration-training/05_SYCL_Migration_OceanFFT/run_dpct_output.sh

@@ -0,0 +1,9 @@
+#!/bin/bash
+source /opt/intel/oneapi/setvars.sh > /dev/null 2>&1
+cd dpct_output/Samples/4_CUDA_Libraries/oceanFFT
+if [ ! -d bin ]; then mkdir bin; fi
+cp -rf ../../../data/* bin/
+icpx -fsycl -fsycl-targets=intel_gpu_pvc -I ../../../Common -I ../../../include *.cpp -qmkl -w
+clear
+if [ $? -eq 0 ]; then ./a.out qatest; fi
+

DirectProgramming/C++SYCL/Jupyter/cuda-to-sycl-migration-training/06_SYCL_Migration_SimpleCUDAGraphs/06_SYCL_Migration_SimpleCUDAGraphs.ipynb

@@ -131,11 +131,13 @@
     "Next, use the SYCLomatic Tool (c2s) to migrate the code; it will store the result in the migration folder `dpct_output`:\n",
     "\n",
     "```\n",
-    "c2s -p compile_commands.json --in-root ../../.. --gen-helper-function\n",
+    "c2s -p compile_commands.json --in-root ../../.. --gen-helper-function --use-experimental-features=logical-group\n",
     "```\n",
     "\n",
     "The `--gen-helper-function` option will copy the SYCLomatic helper header files to output directory.\n",
     "\n",
+    "The `-use-experimental-features=logical-group` option is needed since this CUDA example is using CUDA cooperative groups, and SYCLomatic will migrate this using a experimemtal feature currently.\n",
+    "\n",
     "The `--in-root` option will specify the path for all the common include files for the CUDA project.\n",
     "\n",
     "This command should migrate the CUDA source to the C++ SYCL source in a folder named `dpct_output` by default, and the folder will have the C++ SYCL source along with any dependencies from the `Common` folder,\n",
@@ -147,7 +149,7 @@
   },
   {
    "cell_type": "markdown",
-   "id": "5122ddf2-6bcb-4ec7-8daf-18273d48d8a6",
+   "id": "f394c88b-9fb6-4160-b320-1e7491fa7139",
    "metadata": {},
    "source": [
     "## Analyze, Compile and Run the migrated SYCL source\n",
@@ -171,12 +173,30 @@
     "\n",
     "To compile the migrated SYCL code we can use the following command:\n",
     "```\n",
-    "icpx -fsycl -fsycl-targets=intel_gpu_pvc -I ../../../Common -I ../../../include *.cpp -pthread\n",
+    "icpx -fsycl -fsycl-targets=intel_gpu_pvc -I ../../../Common -I ../../../include *.cpp\n",
     "```\n",
     "\n",
     "There may be compile errors based on whether all of the CUDA code was migrated to SYCL or not. The migrated code may also include comments with warning messages, which could help make it easier to fix the errors. These errors have to be manually fixed to get the code to compile.\n",
     "\n",
-    "\n",
+    "#### Build and Run\n",
+    "Select the cell below and click run ▶ to compile and execute the code (expect to see errors):"
+   ]
+  },
+  {
+   "cell_type": "code",
+   "execution_count": null,
+   "id": "51e1df07-401e-4842-95c8-0e753bd41df5",
+   "metadata": {},
+   "outputs": [],
+   "source": [
+    "! ./q.sh run_dpct_output.sh"
+   ]
+  },
+  {
+   "cell_type": "markdown",
+   "id": "8a58a333-ceee-4bea-ba9c-f0982b49449c",
+   "metadata": {},
+   "source": [
     "### Fixing unmigrated SYCL code\n",
     "\n",
     "The manual migration of CUDA Graph API calls to SYCL can be done using two separate approaches,\n",
@@ -243,7 +263,7 @@
     "\n",
     "For more information on memory operations refer [here](https://github.com/taskflow/taskflow/blob/master/taskflow/sycl/syclflow.hpp).\r\n",
     "\n",
-    "##### Option 2 (using SYCL Graph):\n", 
+    "##### Option 2 (using SYCL Graph):\n",
     "Similar to memcpy node, memset operation can also be included as a node through the command graph add method\n",
     "\n",
     "```\n",
@@ -258,7 +278,7 @@
     "                         nodeDependencies.size(), &kernelNodeParams);\n",
     "```\n",
     "\n",
-    "##### Option 1 (using Taskflow): \n", 
+    "##### Option 1 (using Taskflow): \n",
     "The tf::syclFlow::on creates a task to launch the given command group function object and tf::syclFlow::parallel_for creates a kernel task from a parallel_for method through the handler object associated with a command group. The SYCL runtime schedules command group function objects from an out-of-order queue and constructs a task graph based on submitted events.\n",
     "\n",
     "```\n",
@@ -272,7 +292,7 @@
     "    }).name(\"reduce_kernel\");\n",
     "```\n",
     "\n",
-    "##### Option 2 (using SYCL Graph):\n", 
+    "##### Option 2 (using SYCL Graph):\n",
     "Kernel operations are also included as a node through the command graph `add` method. These commands are captured into the graph and executed asynchronously when the graph is submitted to a queue. The `property::node::depends_on` property can be passed here with a list of nodes to create dependency edges on.\n",
     "\n",
     "```\n",
@@ -296,7 +316,7 @@
     "cudaGraphAddHostNode(&hostNode, graph, nodeDependencies.data(),   nodeDependencies.size(), &hostParams);\n",
     "```\n",
     "\n",
-    "##### Option 1 (using Taskflow): \n", 
+    "##### Option 1 (using Taskflow): \n",
     "The tf::syclFlow doesn’t have a host method to run the callable on the host, instead, we can achieve this by creating a subflow graph since Taskflow supports dynamic tasking and runs the callable on the host.\n",
     "\n",
     "```\n",
@@ -313,7 +333,7 @@
     "cudaGraphGetNodes(graph, nodes, &numNodes);\n",
     "```\n",
     "\n",
-    "##### Option 1 (using Taskflow): \n", 
+    "##### Option 1 (using Taskflow): \n",
     "CUDA graph nodes are equivalent to SYCL tasks, both tf::Taskflow and tf::syclFlow classes include num_tasks() function to query the total number of tasks.\n",
     "\n",
     "```\n",
@@ -335,7 +355,7 @@
     "\n",
     "The inputVec_h2d and outputVec_memset tasks run parallelly followed by the reduce_kernel task.\n",
     "\n",
-    "##### Option 2 (using SYCL Graph):\n", 
+    "##### Option 2 (using SYCL Graph):\n",
     "After all the operations are added as a node the graph is finalized using `finalize()` so that no more nodes can be added and creates an executable graph that can be submitted for execution\n",
     "```\n",
     "auto exec_graph = graph.finalize();\r\n",
@@ -348,7 +368,7 @@
     "cudaGraphClone(&clonedGraph, graph);\n",
     "```\n",
     "\n",
-    "##### Option 1 (using Taskflow): \n", 
+    "##### Option 1 (using Taskflow): \n",
     "In SYCL, no clone function is available as Taskflow graph objects are move-only. We can use the std::move() function as shown below to achieve functionality.\n",
     "\n",
     "```\n",
@@ -362,14 +382,14 @@
     "for (int i = 0; i < GRAPH_LAUNCH_ITERATIONS; i++) {\r\n",
     "  cudaGraphLaunch(graphExec, streamForGraph); }\r\n",
     "```\n",
-    "##### Option 1 (using Taskflow): \n", 
+    "##### Option 1 (using Taskflow): \n",
     "A taskflow graph can be run once or multiple times using an executor. run_n() will run the taskflow the number of times specified by the second argument.\n",
     "\n",
     "```\n",
     "exe.run_n(tflow, GRAPH_LAUNCH_ITERATIONS).wait();\n",
     "```\n",
     "\n",
-    "##### Option 2 (using SYCL Graph):\n", 
+    "##### Option 2 (using SYCL Graph):\n",
     "The graph is submitted in its entirety for execution via `handler::ext_oneapi_graph(graph)`.\n",
     "\n",
     "```\n",
@@ -385,7 +405,7 @@
     "cudaGraphDestroy(graph);\n",
     "```\n",
     "\n",
-    "##### Option 1 (using Taskflow): \n", 
+    "##### Option 1 (using Taskflow): \n",
     "tf::Taskflow class has default destructor operators for both tf::executor and tf::taskflow objects created.\n",
     "\n",
     "```\n",

DirectProgramming/C++SYCL/Jupyter/cuda-to-sycl-migration-training/06_SYCL_Migration_SimpleCUDAGraphs/run_dpct_output.sh

@@ -0,0 +1,7 @@
+#!/bin/bash
+source /opt/intel/oneapi/setvars.sh > /dev/null 2>&1
+cd dpct_output/Samples/3_CUDA_Features/simpleCudaGraphs
+icpx -fsycl -fsycl-targets=intel_gpu_pvc -I ../../../Common -I ../../../include *.cpp -w
+clear
+if [ $? -eq 0 ]; then ./a.out; fi
+

DirectProgramming/C++SYCL/Jupyter/cuda-to-sycl-migration-training/06_SYCL_Migration_SimpleCUDAGraphs/sycl_migrated_option1/CMakeLists.txt

@@ -176,7 +176,7 @@ void syclGraphManual(float *inputVec_h, float *inputVec_d,
       exit(0);
   }
   if(q.get_device().get_info<sycl_ext::info::device::graph_support>()
-      == sycl::ext::oneapi::experimental::info::graph_support_level::unsupported){
+      == sycl::ext::oneapi::experimental::graph_support_level::unsupported){
       printf("sycl graph not supported : %s\nExit\n", 
         q.get_device().get_info<sycl::info::device::name>().c_str());
       exit(0);
@@ -254,7 +254,7 @@ void syclGraphCaptureQueue(float *inputVec_h, float *inputVec_d,
       exit(0);
   }
   if(q.get_device().get_info<sycl_ext::info::device::graph_support>()
-      == sycl::ext::oneapi::experimental::info::graph_support_level::unsupported){
+      == sycl::ext::oneapi::experimental::graph_support_level::unsupported){
       printf("sycl graph not supported : %s\nExit\n", 
         q.get_device().get_info<sycl::info::device::name>().c_str());
       exit(0);