diff --git a/clang/lib/DPCT/RulesLang/RulesLang.cpp b/clang/lib/DPCT/RulesLang/RulesLang.cpp index c2ee40561638..4f016f911617 100644 --- a/clang/lib/DPCT/RulesLang/RulesLang.cpp +++ b/clang/lib/DPCT/RulesLang/RulesLang.cpp @@ -8123,6 +8123,13 @@ void DriverDeviceAPIRule::runRule( auto SecArg = CE->getArg(1); if (auto DRE = dyn_cast(SecArg)) { auto AttributeName = DRE->getNameInfo().getAsString(); + if (AttributeName == + "CU_DEVICE_ATTRIBUTE_VIRTUAL_MEMORY_MANAGEMENT_SUPPORTED" && + !DpctGlobalInfo::useExpVirtualMemory()) { + report(CE->getBeginLoc(), Diagnostics::TRY_EXPERIMENTAL_FEATURE, false, + AttributeName, "--use-experimental-features=virtual_mem"); + return; + } auto Search = MapNames::EnumNamesMap.find(AttributeName); if (Search == MapNames::EnumNamesMap.end()) { report(CE->getBeginLoc(), Diagnostics::NOT_SUPPORTED_PARAMETER, false, diff --git a/clang/test/dpct/driver_device_after11.2.cu b/clang/test/dpct/driver_device_after11.2.cu index 0fee08ec7d2a..0e2cce5e5875 100644 --- a/clang/test/dpct/driver_device_after11.2.cu +++ b/clang/test/dpct/driver_device_after11.2.cu @@ -2,14 +2,18 @@ // UNSUPPORTED: v8.0, v9.0, v9.1, v9.2, v10.0, v10.1, v10.2, v11.0, v11.1 // RUN: dpct --format-range=none -out-root %T/driver_device_after11.2 %s --cuda-include-path="%cuda-path/include" -- -std=c++14 -x cuda --cuda-host-only // RUN: FileCheck %s --match-full-lines --input-file %T/driver_device_after11.2/driver_device_after11.2.dp.cpp -// RUN: %if build_lit %{icpx -c -fsycl %T/driver_device_after11.2/driver_device_after11.2.dp.cpp -o %T/driver_device_after11.2/driver_device_after11.2.dp.o %} +// RUN: %if build_lit %{icpx -c -fsycl -DBUILD_TEST %T/driver_device_after11.2/driver_device_after11.2.dp.cpp -o %T/driver_device_after11.2/driver_device_after11.2.dp.o %} #include int main() { int i; CUdevice d; - // CHECK: i = dpct::get_device(d).has(sycl::aspect::ext_oneapi_virtual_mem); +#ifndef BUILD_TEST + // CHECK: /* + // CHECK-NEXT: DPCT1119:{{[0-9]+}}: Migration of CU_DEVICE_ATTRIBUTE_VIRTUAL_MEMORY_MANAGEMENT_SUPPORTED is not supported, please try to remigrate with option: --use-experimental-features=virtual_mem. + // CHECK-NEXT: */ cuDeviceGetAttribute(&i, CU_DEVICE_ATTRIBUTE_VIRTUAL_MEMORY_MANAGEMENT_SUPPORTED, d); +#endif return 0; } diff --git a/clang/test/dpct/driver_device_virtual_mem_after11.2.cu b/clang/test/dpct/driver_device_virtual_mem_after11.2.cu new file mode 100644 index 000000000000..02f4ff8fb45c --- /dev/null +++ b/clang/test/dpct/driver_device_virtual_mem_after11.2.cu @@ -0,0 +1,15 @@ +// UNSUPPORTED: cuda-8.0, cuda-9.0, cuda-9.1, cuda-9.2, cuda-10.0, cuda-10.1, cuda-10.2, cuda-11.0, cuda-11.1 +// UNSUPPORTED: v8.0, v9.0, v9.1, v9.2, v10.0, v10.1, v10.2, v11.0, v11.1 +// RUN: dpct --format-range=none --use-experimental-features=virtual_mem -out-root %T/driver_device_virtual_mem_after11.2 %s --cuda-include-path="%cuda-path/include" -- -std=c++14 -x cuda --cuda-host-only +// RUN: FileCheck %s --match-full-lines --input-file %T/driver_device_virtual_mem_after11.2/driver_device_virtual_mem_after11.2.dp.cpp +// RUN: %if build_lit %{icpx -c -fsycl %T/driver_device_virtual_mem_after11.2/driver_device_virtual_mem_after11.2.dp.cpp -o %T/driver_device_virtual_mem_after11.2/driver_device_virtual_mem_after11.2.dp.o %} + +#include + +int main() { + int i; + CUdevice d; + // CHECK: i = dpct::get_device(d).has(sycl::aspect::ext_oneapi_virtual_mem); + cuDeviceGetAttribute(&i, CU_DEVICE_ATTRIBUTE_VIRTUAL_MEMORY_MANAGEMENT_SUPPORTED, d); + return 0; +}