From patchwork Fri Mar 24 16:35:15 2023 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 8bit X-Patchwork-Submitter: Thomas Schwinge X-Patchwork-Id: 74634 Return-Path: Delivered-To: ouuuleilei@gmail.com Received: by 2002:a59:b0ea:0:b0:3b6:4342:cba0 with SMTP id b10csp759179vqo; Fri, 24 Mar 2023 09:36:04 -0700 (PDT) X-Google-Smtp-Source: AKy350aIRc9yDnTP7YoP+UnnoOFygNB8YuKXGtV6vpGs+3hPbRs0wkJwqqttLlmSMy2PUyVjwqPA X-Received: by 2002:a17:906:3da:b0:933:2f77:ca78 with SMTP id c26-20020a17090603da00b009332f77ca78mr3478032eja.28.1679675764106; Fri, 24 Mar 2023 09:36:04 -0700 (PDT) ARC-Seal: i=1; a=rsa-sha256; t=1679675764; cv=none; d=google.com; s=arc-20160816; b=ozpyIm6rQy2tgCkIb2+o5G78hSioz7u/sHFJbjlJxEZ6gLtnxwXOxICC1LlWn+8SGq BlKdm8YTOafg2jod4lwJMpMtjvny3qRU3H7WghjmbhAxXJK7/mDOcNH1lqQbobBwFZaN Y4Q7C33EPg2lNldmMnTgCKlQ4KJOcAPla3XKT/TAXB48VkRktvDDwolgQBdtAcs0n+CA kcnrEDteujNWNn3T4rCgfPav1ruXHYyWDsHu2oe+u4nqUW5XweT2BbZBynMVUPHyeIW0 7QpsIzCwGMRcnb239udZcfy075Du2Z2YXJg4KgyiFePhxtjNcOBFxvrT9tTUubwRsYUF dp5Q== ARC-Message-Signature: i=1; a=rsa-sha256; c=relaxed/relaxed; d=google.com; s=arc-20160816; h=sender:errors-to:list-subscribe:list-help:list-post:list-archive :list-unsubscribe:list-id:precedence:mime-version:message-id:date :user-agent:subject:cc:to:from:ironport-sdr:dmarc-filter :delivered-to; bh=MZR5QQEAMtfXNoZ4WdgfbnYDtvP4fzf1sKjrcksLtBA=; b=fI2GMhZHgNEbM3F+jUKeMeC28DW/uJ5E82sQrbTrk4D52E7wtugdwrY81cR29p/xHX nUUBBTECYWVBLg3eFsrvmiUJdPTkOisMEDksJYHywLBVzOHH/5/aqcWNcPgehAzG/SHu XMzXgk2uxeMzlJojA7KAiHi8ZRZRF+8TVoZ8+pZfZ1pQQa+l+CyPIQpjSD4/bOZ9tIQX +nUuSGR8zs4dm+0FGILXBDU+X534vDYelO+w6fk/NvLmIcLM/SDmELxvobMAxeTWGHC+ m3z7J6EhEN7SYCNt3TmFpe85sXS3j3637W+q1vTOcWbk0Bu130PuewUNDcmMt+X24aY7 8HAg== ARC-Authentication-Results: i=1; mx.google.com; spf=pass (google.com: domain of gcc-patches-bounces+ouuuleilei=gmail.com@gcc.gnu.org designates 2620:52:3:1:0:246e:9693:128c as permitted sender) smtp.mailfrom="gcc-patches-bounces+ouuuleilei=gmail.com@gcc.gnu.org" Received: from sourceware.org (server2.sourceware.org. [2620:52:3:1:0:246e:9693:128c]) by mx.google.com with ESMTPS id lg9-20020a170906f88900b0091fff42ea83si23541471ejb.273.2023.03.24.09.36.03 for (version=TLS1_3 cipher=TLS_AES_256_GCM_SHA384 bits=256/256); Fri, 24 Mar 2023 09:36:04 -0700 (PDT) Received-SPF: pass (google.com: domain of gcc-patches-bounces+ouuuleilei=gmail.com@gcc.gnu.org designates 2620:52:3:1:0:246e:9693:128c as permitted sender) client-ip=2620:52:3:1:0:246e:9693:128c; Authentication-Results: mx.google.com; spf=pass (google.com: domain of gcc-patches-bounces+ouuuleilei=gmail.com@gcc.gnu.org designates 2620:52:3:1:0:246e:9693:128c as permitted sender) smtp.mailfrom="gcc-patches-bounces+ouuuleilei=gmail.com@gcc.gnu.org" Received: from server2.sourceware.org (localhost [IPv6:::1]) by sourceware.org (Postfix) with ESMTP id 92C343858D28 for ; Fri, 24 Mar 2023 16:36:02 +0000 (GMT) X-Original-To: gcc-patches@gcc.gnu.org Delivered-To: gcc-patches@gcc.gnu.org Received: from esa3.mentor.iphmx.com (esa3.mentor.iphmx.com [68.232.137.180]) by sourceware.org (Postfix) with ESMTPS id 711EC3858D1E for ; Fri, 24 Mar 2023 16:35:22 +0000 (GMT) DMARC-Filter: OpenDMARC Filter v1.4.2 sourceware.org 711EC3858D1E Authentication-Results: sourceware.org; dmarc=none (p=none dis=none) header.from=codesourcery.com Authentication-Results: sourceware.org; spf=pass smtp.mailfrom=mentor.com X-IronPort-AV: E=Sophos;i="5.98,288,1673942400"; d="scan'208,223";a="281704" Received: from orw-gwy-02-in.mentorg.com ([192.94.38.167]) by esa3.mentor.iphmx.com with ESMTP; 24 Mar 2023 08:35:21 -0800 IronPort-SDR: 6fP/iR1WOgIuS6ohHW7bxLF9j2ctEfQJHgxNMlg12GrQCD/GUClEqecmEjLfVPdCFTW6YY5Usv 6D/voEQukcXlraw0WiIQjFZKX4PT5UmzukFpS9ZIJQnSHE0IVhW/bYpa+JvpwFAplkwgdmgVUs qIeIBvjNRfUEhkanaxiTGO4JN58t407g2uODFpyiv13Mhnb4Xtm00nPnEQJg1XZi6eG0EY5R5M EfMosDoxWvFfIW3fN2iXxExI8LgLD8qyuqNe2sfBh/2dRos40hjB3a7qkBbWi+2NbcRuRscEI/ 78g= From: Thomas Schwinge To: CC: Andrew Stubbs Subject: [og12] Add 'libgomp.c/alloc-ompx_host_mem_alloc-1.c' User-Agent: Notmuch/0.29.3+94~g74c3f1b (https://notmuchmail.org) Emacs/28.2 (x86_64-pc-linux-gnu) Date: Fri, 24 Mar 2023 17:35:15 +0100 Message-ID: <87zg8285vg.fsf@euler.schwinge.homeip.net> MIME-Version: 1.0 X-Originating-IP: [137.202.0.90] X-ClientProxiedBy: svr-ies-mbx-10.mgc.mentorg.com (139.181.222.10) To svr-ies-mbx-10.mgc.mentorg.com (139.181.222.10) X-Spam-Status: No, score=-11.9 required=5.0 tests=BAYES_00, GIT_PATCH_0, HEADER_FROM_DIFFERENT_DOMAINS, KAM_DMARC_STATUS, SPF_HELO_PASS, SPF_PASS, TXREP autolearn=ham autolearn_force=no version=3.4.6 X-Spam-Checker-Version: SpamAssassin 3.4.6 (2021-04-09) on server2.sourceware.org X-BeenThere: gcc-patches@gcc.gnu.org X-Mailman-Version: 2.1.29 Precedence: list List-Id: Gcc-patches mailing list List-Unsubscribe: , List-Archive: List-Post: List-Help: List-Subscribe: , Errors-To: gcc-patches-bounces+ouuuleilei=gmail.com@gcc.gnu.org Sender: "Gcc-patches" X-getmail-retrieved-from-mailbox: =?utf-8?q?INBOX?= X-GMAIL-THRID: =?utf-8?q?1761267693797800015?= X-GMAIL-MSGID: =?utf-8?q?1761267693797800015?= Hi! This had fallen out of some earlier work of mine; I've now pushed to devel/omp/gcc-12 branch commit ae2dca26602678f8b70e22da1bce8302c0751b75 "Add 'libgomp.c/alloc-ompx_host_mem_alloc-1.c'", see attached. Grüße Thomas ----------------- Siemens Electronic Design Automation GmbH; Anschrift: Arnulfstraße 201, 80634 München; Gesellschaft mit beschränkter Haftung; Geschäftsführer: Thomas Heurung, Frank Thürauf; Sitz der Gesellschaft: München; Registergericht München, HRB 106955 From ae2dca26602678f8b70e22da1bce8302c0751b75 Mon Sep 17 00:00:00 2001 From: Thomas Schwinge Date: Wed, 15 Feb 2023 11:27:55 +0100 Subject: [PATCH] Add 'libgomp.c/alloc-ompx_host_mem_alloc-1.c' OpenMP 'ompx_host_mem_alloc' is available for host and nvptx offloading as of og12 commit 84914e197d91a67b3d27db0e4c69a433462983a5 "openmp, nvptx: ompx_unified_shared_mem_alloc", and for GCN offloading as of og12 commit c77c45a641fedc3fe770e909cc010fb1735bdbbd "amdgcn, libgomp: low-latency allocator". libgomp/ * testsuite/libgomp.c/alloc-ompx_host_mem_alloc-1.c: New. --- libgomp/ChangeLog.omp | 2 + .../libgomp.c/alloc-ompx_host_mem_alloc-1.c | 77 +++++++++++++++++++ 2 files changed, 79 insertions(+) create mode 100644 libgomp/testsuite/libgomp.c/alloc-ompx_host_mem_alloc-1.c diff --git a/libgomp/ChangeLog.omp b/libgomp/ChangeLog.omp index ef957e3d2d8..6b816e46cd2 100644 --- a/libgomp/ChangeLog.omp +++ b/libgomp/ChangeLog.omp @@ -1,5 +1,7 @@ 2023-03-24 Thomas Schwinge + * testsuite/libgomp.c/alloc-ompx_host_mem_alloc-1.c: New. + * config/gcn/allocator.c (gcn_memspace_free): Explicitly handle 'memspace == ompx_host_mem_space'. diff --git a/libgomp/testsuite/libgomp.c/alloc-ompx_host_mem_alloc-1.c b/libgomp/testsuite/libgomp.c/alloc-ompx_host_mem_alloc-1.c new file mode 100644 index 00000000000..683b7aff3d4 --- /dev/null +++ b/libgomp/testsuite/libgomp.c/alloc-ompx_host_mem_alloc-1.c @@ -0,0 +1,77 @@ +/* Verify that on the host we can but on a device we cannot allocate 'ompx_host_mem_alloc' memory. */ + +/* { dg-additional-options -DOFFLOAD_DEVICE { target offload_device } } */ + +#include + +#pragma omp requires dynamic_allocators + +int main() +{ +#pragma omp target + { + char *c, *c_; + + c = omp_alloc(1, ompx_host_mem_alloc); +#ifdef OFFLOAD_DEVICE + if (c) + __builtin_abort (); +#else + if (!c) + __builtin_abort (); +#endif + omp_free(c, ompx_host_mem_alloc); + + c = omp_aligned_alloc(128, 256, ompx_host_mem_alloc); +#ifdef OFFLOAD_DEVICE + if (c) + __builtin_abort (); +#else + if (!c) + __builtin_abort (); +#endif + omp_free(c, omp_null_allocator); + + c = omp_calloc(1, 1, ompx_host_mem_alloc); +#ifdef OFFLOAD_DEVICE + if (c) + __builtin_abort (); +#else + if (!c) + __builtin_abort (); +#endif + c_ = omp_realloc(c, 2, ompx_host_mem_alloc, ompx_host_mem_alloc); +#ifdef OFFLOAD_DEVICE + if (c_) + __builtin_abort (); +#else + if (!c_) + __builtin_abort (); +#endif + c = omp_realloc(c_, 0, ompx_host_mem_alloc, ompx_host_mem_alloc); + if (c) + __builtin_abort (); + + c = omp_aligned_calloc(64, 1, 512, ompx_host_mem_alloc); +#ifdef OFFLOAD_DEVICE + if (c) + __builtin_abort (); +#else + if (!c) + __builtin_abort (); +#endif + c_ = omp_realloc(c, 2, c ? omp_null_allocator : ompx_host_mem_alloc, omp_null_allocator); +#ifdef OFFLOAD_DEVICE + if (c_) + __builtin_abort (); +#else + if (!c_) + __builtin_abort (); +#endif + c = omp_realloc(c_, 0, omp_null_allocator, omp_null_allocator); + if (c) + __builtin_abort (); + } + + return 0; +} -- 2.25.1