Skip to content

Implement chipStar support#3

Open
pvelesko wants to merge 819 commits intomainfrom
chipStar
Open

Implement chipStar support#3
pvelesko wants to merge 819 commits intomainfrom
chipStar

Conversation

@pvelesko
Copy link
Copy Markdown
Collaborator

No description provided.

zatkins-dev and others added 30 commits May 6, 2025 16:44
Reworks the stream implementation for `/gpu/hip/gen` to avoid creating and destroying streams on every operator apply.

Updates `hipblas` calls to only sync stream, this matters on MI300A since `hipblas` seems to use an async stream. Avoids a full device sync.

Also makes working vectors come from the `Vector` object delegate to avoid bad ref behavior.
* Fixes for oneAPI-2025

* Neat fix for oneAPI 2025

* style changes
ex - add PETSc BP1+3 and BP2+4 examples
test - fix path to stay in folder
rust - update rust examples to match c/python
Use HTML comments as doc inclusion markers
Add ncomp=4 as common CPU restriction variant
Co-authored-by: Peter Munch <peterrmuench@gmail.com>
ex - use deal.II fe to build libCEED Basis
cpu - remove vectorize warnings for Clang -O3
* doc - improve internal API documentation

* doc - clarify delegation

* doc - update wording for clarity

Co-authored-by: James Wright <james@jameswright.xyz>

* doc - clarify memcheck==valgrind

* doc - clarify shared gpu backends

* doc - clarify l/e/q-vecs

* doc - clarify occa backend status

* doc - more inheritance clarification

* doc - minor claification to QF fields

* minor - remove unneeded variables

* doc - update inheritance language

* doc - update shared description

* doc - more gpu notes

* doc - update fallback description

* doc - clarify fallback is only for pc support

* doc - minor updates to user facing GPU section

* doc - minor reordering of major sections

---------

Co-authored-by: James Wright <james@jameswright.xyz>
jeremylt and others added 20 commits March 2, 2026 14:17
interface: fix uninitialized use in CeedOperatorMultigridLevelCreateSingle_Core
cov - add missing exclusion markers
Better flags for cov accuracy on Noether
* contributing - draft LLM statement

* contributing - deleniate permitted and disallowed LLM usage

* contributing - grammar

Co-authored-by: Valeria Barra <39932030+valeriabarra@users.noreply.github.com>

* dev - add PR template

* contributing - wording consistency between PR template and CONTRIBUTING.md

* minor - typo

* contributing - refrence sf dora

* minor - fix spelling

Co-authored-by: Yohann <dudouit1@llnl.gov>

---------

Co-authored-by: Valeria Barra <39932030+valeriabarra@users.noreply.github.com>
Co-authored-by: Yohann <dudouit1@llnl.gov>
rust - allow setting ceed OPT flags for rust, mostly for CI/cov
@pvelesko pvelesko force-pushed the chipStar branch 2 times, most recently from d2d8709 to a65eb92 Compare March 18, 2026 06:36
pvelesko added 3 commits April 8, 2026 17:32
Detect HIP platform at build time via hipconfig output:
- __HIP_PLATFORM_SPIRV__ → HIP_LIB_NAME=CHIP (chipStar)
- __HIP_PLATFORM_HCC__/__HIP_PLATFORM_AMD__ → HIP_LIB_NAME=amdhip64

Move ROCM_DIR and HIP_ARCH defaults to the top of the file where other
tool-path defaults live. Use HIP_LIB_NAME in the library detection glob
and in PKG_LIBS. Remove the subst=,, stripping from HIPCONFIG_CPPFLAGS
so flags are passed through unmodified.
When SYCL backends are built, libceed.so must be linked with icpx
(SYCLCXX) rather than g++, and -fsycl must appear in CEED_LDFLAGS
(before object files) so icpx can merge the SYCL fat binary device
sections. Without this, libceed.so lacks NEEDED: libsycl.so.7 and
SYCL kernels fail to load at runtime.
Replace the outer element for-loop with a single element assignment and
guard all memory accesses with if (elem < num_elem). Shared memory
operations (Interp, Grad, etc.) must execute unconditionally across all
threads so __syncthreads() is reached uniformly; only the
load/store steps are guarded. Also guard qfunction calls with
thread-id bounds checks and comment out the pragma unroll that
triggered miscompilation on chipStar's LLVM.
@pvelesko pvelesko force-pushed the chipStar branch 2 times, most recently from 6a5b77b to e8e7c92 Compare April 8, 2026 14:46
pvelesko added 2 commits April 9, 2026 09:46
…r gcc

chipStar's hipconfig -C outputs --offload=spirv64, -nohipwrapperinc, --hip-path=,
and --target= which are clang-only flags.  When CC=gcc is used for .c files (or
CXX != HIPCC for .cpp files), these flags cause build failures.

Add HIPCONFIG_CPPFLAGS_C that filters the clang-only flags and adds an explicit
-I$(ROCM_DIR)/include (since -nohipwrapperinc was suppressing the wrapper that
would have pulled in hip_runtime.h).
Add a generic --env KEY=VAL CLI option (repeatable) to tests/junit.py
that injects environment variables into the test subprocess environment.
The flag mutates os.environ in the parent before mp.Pool is created, so
worker processes inherit the values via the existing init_process()
copy.

Useful for backend-specific runtime knobs that affect test behavior, for
example silencing chipStar runtime informational/warning lines that
JUnit otherwise classifies as test failures:

    make junit BACKENDS='/gpu/hip/ref ...' \
        JUNIT_ARGS='--env CHIP_LOGLEVEL=crit'

No changes to tests/junit_common.py.
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.