Tracker Rusticl on r600
Purpose
Purpose of this thread is to track track efforts on making r600 working with rusticl, track user branches, list related merge requests, conduct tests, report issues in rusticl or r600 code that can affect rusticl, etc.
Disclaimer: I don't pretend to have the skills to implement rusticl for r600 all by myself, but I can provide a lot of testing and feedback.
Context
AMD provided OpenCL support with fglrx for TeraScale 1 starting with HD 4000, TeraScale 2 and TeraScale 3 but fglrx has no compatible maintained kernel anymore since years.
Clover provided incomplete OpenCL support for TeraScale 2 and TeraScale 3 but there is no working image support and LLVM regressions broke TeraScale 2 and broke TeraScale 3.
rusticl can be seen as a new hope for OpenCL support on this hardware.
Status
-
2022-08-02:clinfois working for TeraScale 2 and 3:
https://dl.illwieckz.net/b/mesa/wip/rusticl/r600/20221002-006.rusticl-r600-clinfo.md -
2023-06-19:clinfois broken again. -
2023-11-03:clinfois working again. -
2024-05-14: maybe it's better to disable rusticl for r600 then, until somebody really has time to figure all those things out, not sure. And users can still use clover. -- !28785 (comment 2413609)
No more yet.
Conformance
- Too early to test for it
Merge requests
- !18581 (merged) — Rusticl: reduce kernel launch overhead
- !18962 (merged) — r600: set clear_buffer = u_default_clear_buffer
- !18963 (merged) — r600: info.stage MESA_SHADER_KERNEL as MESA_SHADER_COMPUTE
- !19416 (merged) — r600/sfn: lower uniforms to UBOs
- !19417 (merged) — r600/sfn: Handle load_workgroup_size
- !20252 (merged) — r600/sfn: Add experimental support for rustcl
- !23703 (merged) — gallium: add timer resolution pipe cap
- !24229 (merged) — r600: Add experimental get_compute_state_info
-
#12985 (closed) - Not really a merge-request, but a shared patch to fix
clinfoagain.
Issues
- #10087 — rusticl/r600: querying devices prints “r600 device_uuid output is based on invalid pci bus info”
- #10088 — rusticl/r600: unknown instruction '32 %491 = u2u32 %490' in sfn_instr_alu.cpp
- #10089 — rusticl/r600: GPU lock-up when running cl-mem on TeraScale 2 Cypress
- #10090 — rusticl/r600: cl-mem aborts with “clCreateBuffer (-61)” on TeraScale 2 Turks and Redwood
- #10091 — rusticl/r600: sfn_valuefactory.cpp: r600::VirtualValue* r600::ValueFactory::ssa_src(const nir_def&, int): Assertion `!"" "Source values should always exist"' failed.
Hardware testing table
If you have access to some hardware and are willing to test you can add what kind of hardware you have access to in comments. The table is sorted from newer hardware to older hardware (hardware release date) even if architecture may be older.
| Tester | Brand | Vendor | Name | Code name | Micro arch. | Range | Note | Status |
|---|---|---|---|---|---|---|---|---|
| illwieckz | AMD | Asus | Radeon HD 7450 | RV910, Caicos | TeraScale 2 | low-profile | low-end | untested |
| illwieckz | AMD | Gigabyte | Radeon HD 6990 | RV970, Antilles XT | TeraScale 3 | high-profile | high-end, dual-GPU | untested |
| randrianasulu | AMD | ?? | Radeon HD 6400M/7400M | Seymour (Caicos) | TeraScale 2 | APU | low-end, mobile | untested |
| illwieckz | AMD | XFX | Radeon HD 6970 | RV970, Cayman XT | TeraScale 3 | high-profile | high-end | 🚧️ wip |
| illwieckz | AMD | Asus | Radeon HD 6950 | RV970, Cayman PRO | TeraScale 3 | high profile | high-end | untested |
| illwieckz | AMD | Dell | FirePro 3D V4800 | RV830, Redwood XT GL | TeraScale 2 | low-profile | low-end | untested |
| illwieckz | ATi | Gigabyte | Radeon HD 5870 Eyefinity ⁶ | RV870, Cypress XT | TeraScale 2 | high-profile | high-end | 🚧️ wip |
| illwieckz | ATi | MSI | Radeon HD 5450 | RV810, Cedar | TeraScale 2 | low-profile | low-end | untested |
| illwieckz | ATi | Asus | Radeon HD 5670 | RV830, Redwood XT | TeraScale 2 | high-profile | mid-end | untested |
| illwieckz | ATi | ATi | Radeon HD 5970 | RV870, Hemlock XT | TeraScale 2 | high-profile | high-end, dual-GPU | untested |
| illwieckz | ATi | Sapphire | Radeon HD 4890 Vapor-X | RV790 XT | TeraScale 1 | high-profile | high-end | untested |
| illwieckz | ATi | HIS | Radeon HD 4350 | RV710 | TeraScale 1 | low-profile | low-end | untested |
| illwieckz | ATi | HIS | Radeon HD 4670 IceQ | RV730 XT | TeraScale 1 | high-profile | high-end | untested |
| illwieckz | ATi | Club3D | Radeon HD 4870 X2 | RV770 XT | TeraScale 1 | high-profile | high-end, dual-GPU | untested |
Software testing table
| Software | Status |
|---|---|
| clinfo |
|
| cl-mem | ❌️ crash |
| clpeak | ❌️ crash |
| uCLbench | untested |
| SHOC | untested |
| Parboil | untested |
| ViennaCLBench | untested |
| GeekBench | untested |
| FluidX3D | untested |
| OpenCL-Benchmark | untested |
| Einstein@Home | untested |
| LuxMark 3 luxball | ❌️ crash |
| LuxMark 3 microphone | untested |
| LuxMark 3 hotel | untested |
| LuxMark 4.0alpha food | untested |
| LuxMark 4.0alpha hall | untested |
| LuxCoreRender 2.5 | untested |
| BlendLuxCore 2.6 | untested |
| ffmpeg | untested |
| GIMP | untested |
| Darktable | untested |
| Darktable 4 filmic_rgb | untested |
| LibreOffice Calc | untested |
| Natron openfx-arena | untested |
| OBS Studio | untested |
| Davinci Resolve | untested |