Add more tests

.upstream-tests/test/std/thread/thread.mutex/thread.lock/thread.lock.guard/adopt_lock.pass.cpp

@@ -0,0 +1,64 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2022 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+//
+//===----------------------------------------------------------------------===//
+//
+// UNSUPPORTED: no-threads
+// UNSUPPORTED: pre-sm-70
+
+#define _LIBCUDACXX_INLINE_THREADING
+// <mutex>
+
+// template <class Mutex> class lock_guard;
+
+// lock_guard(mutex_type& m, adopt_lock_t);
+
+#include<cuda/std/mutex>
+#include<cuda/std/cassert>
+#include<cuda/std/chrono>
+
+#include "heterogeneous_thread_handler.h"
+#include "test_macros.h"
+
+STATIC_TEST_GLOBAL_VAR
+cuda::std::mutex m;
+
+typedef cuda::std::chrono::system_clock Clock;
+typedef Clock::time_point time_point;
+typedef Clock::duration duration;
+typedef cuda::std::chrono::milliseconds ms;
+typedef cuda::std::chrono::nanoseconds ns;
+
+__host__ __device__
+void f()
+{
+    time_point t0 = Clock::now();
+    time_point t1;
+    {
+    cuda::std::lock_guard<cuda::std::mutex> lg(m, cuda::std::adopt_lock);
+    t1 = Clock::now();
+    }
+    ns d = t1 - t0 - ms(250);
+    assert(d < ms(200));  // within 200ms
+}
+
+int main(int, char**)
+{
+#ifndef __CUDA_ARCH__
+    cuda_thread_count = 2;
+#endif
+    heterogeneous_thread_handler handler;
+    handler.run_on_first_thread(&cuda::std::mutex::lock, m);
+    handler.syncthreads();
+
+    handler.run_on_second_thread(f);
+    handler.sleep_first_thread(ms(250));
+    handler.run_on_first_thread(&cuda::std::mutex::unlock, m);
+    handler.syncthreads();
+    handler.join_test_thread();
+    return 0;
+}

.upstream-tests/test/std/thread/thread.mutex/thread.lock/thread.lock.guard/assign.compile.fail.cpp

@@ -0,0 +1,30 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2022 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: nvrtc
+// UNSUPPORTED: pre-sm-70
+
+// <mutex>
+
+// template <class Mutex> class lock_guard;
+
+// lock_guard& operator=(lock_guard const&) = delete;
+
+#include<cuda/std/mutex>
+
+int main(int, char**)
+{
+    cuda::std::mutex m0;
+    cuda::std::mutex m1;
+    cuda::std::lock_guard<cuda::std::mutex> lg0(m0);
+    cuda::std::lock_guard<cuda::std::mutex> lg(m1);
+    lg = lg0;
+
+  return 0;
+}

.upstream-tests/test/std/thread/thread.mutex/thread.lock/thread.lock.guard/copy.compile.fail.cpp

@@ -0,0 +1,28 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2022 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: nvrtc
+// UNSUPPORTED: pre-sm-70
+
+// <mutex>
+
+// template <class Mutex> class lock_guard;
+
+// lock_guard(lock_guard const&) = delete;
+
+#include<cuda/std/mutex>
+
+int main(int, char**)
+{
+    cuda::std::mutex m;
+    cuda::std::lock_guard<cuda::std::mutex> lg0(m);
+    cuda::std::lock_guard<cuda::std::mutex> lg(lg0);
+
+  return 0;
+}

.upstream-tests/test/std/thread/thread.mutex/thread.lock/thread.lock.guard/implicit_ctad.pass.cpp

@@ -0,0 +1,33 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2022 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: no-threads
+// UNSUPPORTED: c++98, c++03, c++11, c++14
+// UNSUPPORTED: pre-sm-70
+
+// <mutex>
+
+// lock_guard
+
+// Make sure that the implicitly-generated CTAD works.
+
+#include<cuda/std/mutex>
+
+#include "test_macros.h"
+
+int main(int, char**) {
+  cuda::std::mutex mutex;
+  {
+    cuda::std::lock_guard lock(mutex);
+    ASSERT_SAME_TYPE(decltype(lock), cuda::std::lock_guard<cuda::std::mutex>);
+  }
+
+  return 0;
+}
+

.upstream-tests/test/std/thread/thread.mutex/thread.lock/thread.lock.guard/mutex.fail.cpp

@@ -0,0 +1,28 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2022 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: no-threads
+// UNSUPPORTED: nvrtc
+// UNSUPPORTED: pre-sm-70
+
+// <mutex>
+
+// template <class Mutex> class lock_guard;
+
+// explicit lock_guard(mutex_type& m);
+
+#include<cuda/std/mutex>
+
+int main(int, char**)
+{
+    cuda::std::mutex m;
+    cuda::std::lock_guard<cuda::std::mutex> lg = m; // expected-error{{no viable conversion}}
+
+  return 0;
+}

.upstream-tests/test/std/thread/thread.mutex/thread.lock/thread.lock.guard/mutex.pass.cpp

@@ -0,0 +1,78 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2022 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+//
+//===----------------------------------------------------------------------===//
+//
+// UNSUPPORTED: no-threads
+// UNSUPPORTED: pre-sm-70
+
+// <mutex>
+
+// template <class Mutex> class lock_guard;
+
+// explicit lock_guard(mutex_type& m);
+
+// template<class _Mutex> lock_guard(lock_guard<_Mutex>)
+//     -> lock_guard<_Mutex>;  // C++17
+
+#include<cuda/std/mutex>
+#include<cuda/std/cassert>
+#include<cuda/std/chrono>
+
+#include "heterogeneous_thread_handler.h"
+#include "test_macros.h"
+
+STATIC_TEST_GLOBAL_VAR
+cuda::std::mutex m;
+
+typedef cuda::std::chrono::system_clock Clock;
+typedef Clock::time_point time_point;
+typedef Clock::duration duration;
+typedef cuda::std::chrono::milliseconds ms;
+typedef cuda::std::chrono::nanoseconds ns;
+
+__host__ __device__
+void f()
+{
+    time_point t0 = Clock::now();
+    time_point t1;
+    {
+    cuda::std::lock_guard<cuda::std::mutex> lg(m);
+    t1 = Clock::now();
+    }
+    ns d = t1 - t0 - ms(250);
+    assert(d < ms(200));  // within 200ms
+}
+
+__host__ __device__
+void check_ctad()
+{
+#if TEST_STD_VER >= 17
+    cuda::std::lock_guard lg(m);
+    static_assert((cuda::std::is_same<decltype(lg), cuda::std::lock_guard<cuda::std::mutex>>::value), "" );
+    unused(lg);
+#endif
+}
+
+int main(int, char**)
+{
+#ifndef __CUDA_ARCH__
+    cuda_thread_count = 2;
+#endif
+    heterogeneous_thread_handler handler;
+    handler.run_on_first_thread(&cuda::std::mutex::lock, m);
+    handler.syncthreads();
+
+    handler.run_on_second_thread(f);
+    handler.sleep_first_thread(ms(250));
+    handler.run_on_first_thread(&cuda::std::mutex::unlock, m);
+    handler.syncthreads();
+    handler.join_test_thread();
+
+    handler.run_on_first_thread(check_ctad);
+    return 0;
+}

.upstream-tests/test/std/thread/thread.mutex/thread.lock/thread.lock.guard/types.pass.cpp

@@ -0,0 +1,34 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2022 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+//
+//===----------------------------------------------------------------------===//
+//
+// UNSUPPORTED: no-threads
+// UNSUPPORTED: pre-sm-70
+
+// <mutex>
+
+// template <class Mutex>
+// class lock_guard
+// {
+// public:
+//     typedef Mutex mutex_type;
+//     ...
+// };
+
+#include<cuda/std/mutex>
+#include<cuda/std/type_traits>
+
+#include "test_macros.h"
+
+int main(int, char**)
+{
+    static_assert((cuda::std::is_same<cuda::std::lock_guard<cuda::std::mutex>::mutex_type,
+                   cuda::std::mutex>::value), "");
+
+  return 0;
+}

.upstream-tests/test/std/thread/thread.mutex/thread.lock/thread.lock.scoped/adopt_lock.pass.cpp

@@ -0,0 +1,75 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2022 NVIDIA CORPORATION & AFFILIATES. All rights reserved.
+//
+//===----------------------------------------------------------------------===//
+//
+// UNSUPPORTED: no-threads
+// UNSUPPORTED: c++03, c++11, c++14
+// UNSUPPORTED: pre-sm-70
+
+// <mutex>
+
+// template <class ...Mutex> class scoped_lock;
+
+// scoped_lock(adopt_lock_t, Mutex&...);
+
+#include<cuda/std/mutex>
+#include<cuda/std/cassert>
+#include "test_macros.h"
+
+struct TestMutex {
+    bool locked = false;
+    TestMutex() = default;
+
+    __host__ __device__ void lock() { assert(!locked); locked = true; }
+    __host__ __device__ bool try_lock() { if (locked) return false; locked = true; return true; }
+    __host__ __device__ void unlock() { assert(locked); locked = false; }
+
+    TestMutex(TestMutex const&) = delete;
+    TestMutex& operator=(TestMutex const&) = delete;
+};
+
+int main(int, char**)
+{
+    {
+        using LG = cuda::std::scoped_lock<>;
+        LG lg(cuda::std::adopt_lock);
+    }
+    {
+        TestMutex m1;
+        using LG = cuda::std::scoped_lock<TestMutex>;
+        m1.lock();
+        {
+            LG lg(cuda::std::adopt_lock, m1);
+            assert(m1.locked);
+        }
+        assert(!m1.locked);
+    }
+    {
+        TestMutex m1, m2;
+        using LG = cuda::std::scoped_lock<TestMutex, TestMutex>;
+        m1.lock(); m2.lock();
+        {
+            LG lg(cuda::std::adopt_lock, m1, m2);
+            assert(m1.locked && m2.locked);
+        }
+        assert(!m1.locked && !m2.locked);
+    }
+    {
+        TestMutex m1, m2, m3;
+        using LG = cuda::std::scoped_lock<TestMutex, TestMutex, TestMutex>;
+        m1.lock(); m2.lock(); m3.lock();
+        {
+            LG lg(cuda::std::adopt_lock, m1, m2, m3);
+            assert(m1.locked && m2.locked && m3.locked);
+        }
+        assert(!m1.locked && !m2.locked && !m3.locked);
+    }
+
+
+  return 0;
+}