code: def test_coarse_grain_svm(ctx_factory, use_opaque_style):
      import sys
  
      is_pypy = "__pypy__" in sys.builtin_module_names
  
      ctx = ctx_factory()
      queue = cl.CommandQueue(ctx)
  
      dev = ctx.devices[0]
  
      from pyopencl.characterize import has_coarse_grain_buffer_svm
      from pytest import skip
  
      if not has_coarse_grain_buffer_svm(queue.device):
          skip("device does not support coarse-grain SVM")
  
      if "AMD" in dev.platform.name and dev.type & cl.device_type.CPU:
          pytest.xfail("AMD CPU doesn't do coarse-grain SVM")
      if "AMD" in dev.platform.name and dev.type & cl.device_type.GPU:
          pytest.xfail("AMD GPU crashes on SVM unmap")
      if (
          dev.platform.vendor == "The pocl project"
          and dev.type & cl.device_type.GPU
          and "k40" in dev.name.lower()
      ):
          pytest.xfail("Crashes on K40s via PoCL-CUDA")
  
      dtype = np.dtype(np.float32)
      n = 3000
      if use_opaque_style:
          svm_ary = cl.SVMAllocation(
              ctx,
              n * dtype.itemsize,
              alignment=64,
              flags=cl.svm_mem_flags.READ_WRITE,
          )
      else:
          svm_ary = cl.SVM(
              cl.csvm_empty(ctx, (n,), dtype, alignment=64)
          )
          if not is_pypy:
              # https://bitbucket.org/pypy/numpy/issues/52
              assert isinstance(svm_ary.mem.base, cl.SVMAllocation)
  
      cl.enqueue_svm_memfill(queue, svm_ary, np.zeros((), dtype))
  
      with svm_ary.map_rw(queue) as ary:
          if use_opaque_style:
              ary = ary.view(dtype)
          else:
              assert ary is svm_ary.mem
  
          assert ary.nbytes == n * dtype.itemsize
  
          ary.fill(17)
          orig_ary = ary.copy()
  
      prg = cl.Program(
          ctx,
          """
          __kernel void twice(__global float *a_g)
          {
            a_g[get_global_id(0)] *= 2;
          }
          """,
      ).build()
  
      prg.twice(queue, (n,), None, svm_ary)
  
      if (
          dev.platform.vendor == "The pocl project"
          and dev.type & cl.device_type.GPU
      ):
          # clCreateBuffer from SVM doesn't work yet on GPU pocl
          prg.twice(queue, (n,), None, svm_ary)
      else:
          prg.twice(queue, (n,), None, svm_ary.as_buffer(ctx))
  
      with svm_ary.map_ro(queue) as ary:
          if use_opaque_style:
              ary = ary.view(dtype)
          else:
              assert ary is svm_ary.mem
  
          assert np.array_equal(orig_ary * 4, ary)
  
      new_ary = np.empty_like(orig_ary)
      new_ary.fill(-1)
  
      cl.enqueue_copy(queue, new_ary, svm_ary)
      assert np.array_equal(orig_ary * 4, new_ary)
  
      # {{{ https://github.com/inducer/pyopencl/issues/372
  
      buf_arr = cl.svm_empty(
          ctx, cl.svm_mem_flags.READ_ONLY, 10, np.int32
      )
      out_arr = cl.svm_empty(
          ctx, cl.svm_mem_flags.READ_WRITE, 10, np.int32
      )
  
      svm_buf_arr = cl.SVM(buf_arr)
      svm_out_arr = cl.SVM(out_arr)
      with svm_buf_arr.map_rw(queue) as ary:
          ary.fill(17)
  
      prg_ro = cl.Program(
          ctx,
          r"""
          __kernel void twice_ro(__global int *out_g, __global int *in_g)
          {
            out_g[get_global_id(0)] = 2*in_g[get_global_id(0)];
          }
          """,
      ).build()
  
      prg_ro.twice_ro(
          queue, buf_arr.shape, None, svm_out_arr, svm_buf_arr
      )
  
      with svm_out_arr.map_ro(queue) as ary:
          print(ary)
  
imports: import numpy
  import numpy.linalg
  import pytest
  import pyopencl
  import pyopencl.array
  import pyopencl.cltypes
  import pyopencl.clrandom
  from pyopencl.tools import pytest_generate_tests_for_pyopencl, ImmediateAllocator, DeferredAllocator
  from pyopencl.characterize import get_pocl_version
label_desc: Use image node_id 32 for calling a global function with example usage: test_coarse_grain_svm(ctx_factory, use_opaque_style) without return types