Revert windows-related changes for void *params[] in third_party/nvidia/backend/driver.py (#2939)

So it's more like what other backends have. If this was a change from an
old pull request (in Triton), I guess we can roll it back. @gshimansky
please take a look.

Signed-off-by: Anatoly Myachev <[email protected]>

Author: anmyachev

third_party/nvidia/backend/driver.py

@@ -235,7 +235,7 @@ def format_of(ty):
 #endif
 
 static void _launch(int gridX, int gridY, int gridZ, int num_warps, int num_ctas, int clusterDimX, int clusterDimY, int clusterDimZ, int shared_memory, CUstream stream, CUfunction function, CUdeviceptr global_scratch{', ' + arg_decls if len(arg_decls) > 0 else ''}) {{
-  void *params[] = {{{', '.join(f'&arg{i}' for i in params) if params else 'NULL'}}};
+  void *params[] = {{ {', '.join(params)} }};
   if (gridX*gridY*gridZ > 0) {{
     if (num_ctas == 1) {{
       CUDA_CHECK(cuLaunchKernel(function, gridX, gridY, gridZ, 32*num_warps, 1, 1, shared_memory, stream, params, 0));